NVCC compile to ptx using CMAKE's cuda_compile_ptx

Question

I have the simple kernel in file kernel.cu

__global__ void add1( double * pi, double c ) 
{
    *pi += c;
}

and can easily compile it to a ptx file kernel.ptx with:

nvcc -ptx kernel.cu

now, I wanted to reproduce the same behaviour using cmake with the following CMakeLists.txt:

cmake_minimum_required(VERSION 2.8)
project(cmake_ptx)
find_package(CUDA REQUIRED)
cuda_compile_ptx(
  test
  kernel.cu
)

but when I type

cmake . && make

no ptx file is built. How can I reproduce the behavior of the above nvcc command using cmake? In particular, how to:

add the standalone compilation of ptx files to a target (e.g. all): I noticed that when there is another cuda_add_executable inside the same CMakeFile, it also builds the ptx file, otherwise not.
name the file like the source file but with .ptx instead of .cu ending: related to this question: How do I change the output filename of cuda_compile_ptx in CMake?

score 5 · Accepted Answer · answered Oct 05 '14 at 07:46

5

cuda_compile_ptx creates only rules to generate files, but doesn't add them to any target. You need to add custom target, that depends on ptx files:

cmake_minimum_required(VERSION 2.8)
project(cmake_ptx)
find_package(CUDA REQUIRED)
cuda_compile_ptx(
  cuda_ptx_files
  kernel.cu
)
add_custom_target(ptx ALL
    DEPENDS ${cuda_ptx_files} kernel.cu
    SOURCES kernel.cu)

Now if you run make or make ptx, it will generate the ptx files.

answered Oct 05 '14 at 07:46

vinograd47

6,320
28
30

great, that works, but it gives me a file called `cuda_compile_ptx_generated_kernel.cu.ptx` because of this line in FindCUDA: (see [thread](http://stackoverflow.com/questions/10135795/how-do-i-change-the-output-filename-of-cuda-compile-ptx-in-cmake) I mentioned above) `set(generated_file_basename "${cuda_target}_generated_${basename}.ptx")` since I don't understand how to hack the FindCUDA File just using my CMakeList, is there a way to rename the output stored in ${cuda_ptx_files} into the format I want using CMake? – adrelino Oct 05 '14 at 08:33
You can rename it with `add_custom_command`: `add_custom_command(OUTPUT ${CMAKE_BINARY_DIR}/my_ptx.ptx COMMAND ${CMAKE_COMMAND} -E copy ${cuda_ptx_files} ${CMAKE_BINARY_DIR}/my_ptx.ptx DEPENDS ${cuda_ptx_files})` and use `${CMAKE_BINARY_DIR}/my_ptx.ptx` in `add_custom_target`. – vinograd47 Oct 05 '14 at 09:12
@jet47: Can I make that copying generic, so that the .ptx files will be generated where .o files would be generated? Or in some other place, but with the same filenames as the .cu's? – einpoklum Feb 28 '16 at 22:17

score 1 · Answer 2 · answered Oct 05 '14 at 18:35

Just for reference, this is what worked for my project setup (I have a src/ and an include/ folder), using the tips from the accepted answer:

cmake_minimum_required(VERSION 2.8)

project(cmake_ptx)

find_package(CUDA REQUIRED)
include_directories(include/)

cuda_compile_ptx(
  cuda_ptx_files
  src/common_kernels.cu
  OPTIONS -DCUDA_MATLAB
)
add_custom_command(OUTPUT ${CMAKE_BINARY_DIR}/common_kernels.ptx COMMAND ${CMAKE_COMMAND} -E rename ${cuda_ptx_files} ${CMAKE_BINARY_DIR}/common_kernels.ptx DEPENDS ${cuda_ptx_files}) 
add_custom_target(ptx ALL
    DEPENDS ${CMAKE_BINARY_DIR}/common_kernels.ptx src/common_kernels.cu
    SOURCES src/common_kernels.cu
)

This gave me exactly the same output as invoking:

nvcc -ptx src/common_kernels.cu -I include/ -DCUDA_MATLAB

NVCC compile to ptx using CMAKE's cuda_compile_ptx

2 Answers2