CMake to generate a MSVC CUDA project that targets

My PC has a GTX 580 (compute capability 2.0).

I want to compile a CUDA source that uses dynamic parallelism, a feature introduced in compute capability 3.5.

I know I will not be able to run the program on my GPU, however, it should be possible to compile this code on my machine. I'm assuming this because I can compile with no problems the CUDA samples that use 3.5 capability. These samples come with Visual Studio projects that were "manually generated" (I guess).

I believe my problem is with CMake. I'm using CMake to generate a Visual Studio 2012 project.

My first CMakeLists.txt looked like this:

PROJECT(sample-cuda-tests)

FIND_PACKAGE(CUDA REQUIRED)

INCLUDE_DIRECTORIES(${CUDA_INCLUDE_DIRS})
INCLUDE_DIRECTORIES(${CMAKE_CURRENT_SOURCE_DIR}/include)

FILE(GLOB_RECURSE includes ${CMAKE_CURRENT_SOURCE_DIR}/include/*.h )
FILE(GLOB_RECURSE sources ${CMAKE_CURRENT_SOURCE_DIR}/src/*.cc ${CMAKE_CURRENT_SOURCE_DIR}/src/*.cu )

CUDA_ADD_EXECUTABLE(sample-cuda-tests ${includes} ${sources})
TARGET_LINK_LIBRARIES(sample-cuda-tests ${CUDA_LIBRARIES})

Then, when compiling with the generated Visual Studio 2012 project, I got a warning followed by an error:

warning : The 'compute_10' and 'sm_10' architectures are deprecated, and may be removed in a future release.

error : calling a __global__ function from a __global__ function is only allowed on the compute_35 architecture or above

What was expected. Then I added

list(APPEND CUDA_NVCC_FLAGS -gencode arch=compute_35,code=sm_35)

to the CMakeLists. The warning disappeared, but I got:

error : kernel launch from __device__ or __global__ functions requires separate compilation mode

Ok. So I added to the CMakeLists:

set(CUDA_SEPARABLE_COMPILATION ON)

...and received this:

fatal error : nvcc supports '--relocatable-device-code=true (-rdc=true)', '--device-c (-dc)', and '--device-link (-dlink)' only when targeting sm_20 or higher

What is weird because I thought I was targeting sm_35 (higher than sm_20).

Later I discovered I can set some options directly in CUDA_ADD_EXECUTABLE command. So I removed the line that was appending values to CUDA_NVCC_FLAGS and changed CUDA_ADD_EXECUTABLE command to:

CUDA_ADD_EXECUTABLE(sample-cuda-tests ${includes} ${sources} OPTIONS -gencode arch=compute_35,code=sm_35)

What I got was:

C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.0\bin\crt\link.stub : fatal error C1083: Cannot open compiler generated file: 'C:/Users/sms/Desktop/sample-cuda-tests/CMakeFiles/sample-cuda-tests.dir/Debug/sample-cuda-tests_intermediate_link.obj': No such file or directory

No idea where to go now. Appreciate any help.

I'm using CUDA SDK 6.0 on Windows 7.

回答1:

As of CMake 3.1.0, CMake script misses creating a directory to put the intermediate file in. Add the following snippet in the FindCUDA.cmake

get_filename_component(output_file_path "${output_file}" PATH)
add_custom_command(
  TARGET ${cuda_target}
  PRE_LINK
  COMMAND ${CMAKE_COMMAND} -E make_directory ${output_file_path}
)

right before

if (do_obj_build_rule)

in function CUDA_LINK_SEPARABLE_COMPILATION_OBJECTS

回答2:

Turned out to be a bug on FindCUDA.cmake.

When setting CUDA_SEPARABLE_COMPILATION to ON, if .cu files are not in the same folder of CMakeLists.txt, intermediate linkage objects are generated in the wrong folder, causing a compilation error that, on Visual Studio, looks like this:

Cannot open compiler generated file: 'project_path/CMakeFiles/project_name/Debug/project_name_intermediate_link.obj': No such file or directory.

I've opened an issue in CMake bug tracker: http://public.kitware.com/Bug/view.php?id=15016 (the bug is better described there)