CUDA front-end (CUDA to LLVM IR)

Project Infrastructure LLVM Dev List Archives
1

Hi,

I wanted to ask whether there is ongoing effort (or an already established tool) that enables to convert CUDA kernels (that uses CUDA specific intrinsics, e.g., threadId.x, __syncthreads(), …) to LLVM IR. I am aware that I can do this for OpenCL with the help of libclc but I can not find something similar for CUDA.

Thanks

2

A tool of this kind here: https://github.com/apc-llc/nvcc-llvm-ir

3

There’s been a discussion about it on the Clang mailing list (cfe-dev@) just this week - please take a look there

Eli

4

There is no any "License" file in your GitHub repository .
Will it be included ?

Thank you very much .

Mehmet Erol Sanliturk

5

Hi Eli,

Could you please point me to the exact thread ? I skimmed through the titles sent on April and nothing looked related.

Thanks a lot

6

Added

7

Added

Without explicit license information , works can not be used by the others .

Thank you very much .

Mehmet Erol Sanliturk

8

Hi Eli,

Could you please point me to the exact thread ? I skimmed through the
titles sent on April and nothing looked related.

Thanks a lot

[cfe-dev] Parallelism TS implementation and feasibility of GPU execution
policies

Eli

9

Hi Dmitry,

I tried using the tool you pointed me to and run into a problem, you may be able to help.

Here is what I did:

  • I installed (i.e., downloaded and built) LLVM 3.2 as it seems that this is the version that is compatible with the tool.

  • I changed the hard-coded paths in the nvcc-llvm-ir/makefile as follows (changes are in bold)

libcicc.so: cicc.cpp
g++ -g -O3 -D__STDC_LIMIT_MACROS -D__STDC_CONSTANT_MACROS -I/$LLVM_BUILD_DIR/include/ -I/$CUDA_5.5_BUILD_DIR/nvvm/include/ -fPIC $< -shared -o $@ -ldl -L/$LLVM_BUILD_DIR/Release+Asserts/lib/ -Wl,–start-group -lLLVMCore -lLLVMSupport -lLLVMipo -lLLVMipa -lLLVMAnalysis -lLLVMTarget -lLLVMScalarOpts -lLLVMTransformUtils -lLLVMInstCombine -Wl,–end-group -lpthread

libnvcc.so: nvcc.cpp
g++ -g -O3 -I/$CUDA_5.5_BUILD_DIR/nvvm/include/ -fPIC $< -shared -o $@ -ldl

-I noticed that some of the required header files from LLVM exists in -I/$LLVM_SRC_DIR/include/ and others are auto-generated into -I/$LLVM_BUILD_DIR/include/ so I had to manually move them to one folder for the make to compiler properly (I moved them to the build directory).

-It compiles fine and both libcicc.so and libnvcc.so are generated.

  • Finally, when I try to run ‘CICC_MODIFY_UNOPT_MODULE=1 LD_PRELOAD=./libnvcc.so nvcc -arch=sm_23 test1.cu -c -keep’ , it gives a Segmentation fault.

In fact doing “export LD_PRELOAD=./libnvcc.so” was enough to make the terminal Segfault with any other command !!! (e.g., it segfault when simply doing “ls” if LD_PRELOAD is exported)

Any idea what is the problem ?

This is my OS info:

Distributor ID: SUSE LINUX
Description: openSUSE 11.4 (x86_64)
Release: 11.4
Codename: Celadon