使用cmake为vs2013配置cuda项目出现“invalid device function”错误

Posted 2023-03-23

【中文标题】使用cmake为vs2013配置cuda项目出现“invalid device function”错误

【英文标题】：Use cmake to configure cuda project for vs2013 and get "invalid device function" error

【发布时间】：2016-01-12 14:21:42

【问题描述】：

我使用 cmake gui 工具在 vs2013 中配置我的 cuda 项目。 CMakeLists.txt 如下：

project(CUDA_PART)
# required cmake version
cmake_minimum_required(VERSION 3.0)

include_directories($CUDA_PART_SOURCE_DIR/common)
# packages
find_package(CUDA REQUIRED)

# nvcc flags
set(CUDA_NVCC_FLAGS -gencode arch=compute_20,code=sm_20;-G;-g)
set(CUDA_VERBOSE_BUILD ON)

#FILE(GLOB SOURCES "*.cu" "*.cpp" "*.c" "*.h")
CUDA_ADD_EXECUTABLE(CUDA_PART hist_gpu_shmem_atomics.cu)

.cu 文件来自 Cuda by example 源代码hist_gpu_shmem_atomics.cu

有两个问题：

histo_kernel <<<blocks * 2, 256 >>>(dev_buffer, SIZE, dev_histo);行之后出现“无效设备功能”错误。

使用CUDA调试工具调试时，无法触发设备代码中的断点。

但是当我在 Visual Studio 2013 的 cuda 项目 Temple 中创建一个具有相同代码的项目时，它可以正常工作！

那么，CMakeLists.txt 有什么问题吗？

操作系统：Win7 64bit；GPU：GTX960；CUDA：CUDA 7.5；VS：2013（和也是 2010 年）

---

当我使用 vs2013 中的“代码生成”设置如下：

CUDA_NVCC_FLAGES 结果是-gencode=arch=compute_20,code=\"sm_20,compute_20\" 等于：

-gencode=arch=compute_20,code=sm_20 \
-gencode=arch=compute_20,code=compute_20 

所以，我猜它会生成两个版本的机器代码：第一个（SASS）具有虚拟和真实架构，第二个（PTX）只有虚拟架构。由于我的 GTX960 是 cc5.2 设备，所以它选择了第二个（PTX）并将其转换为合适的 SASS。

【问题讨论】：

什么是“CUDA调试工具”？

在VS2013中使用CUDA项目模板编译时使用了哪些nvcc标志？

@Drop 我在vs2013中使用Nsight作为cuda调试工具。

@m.s. nvcc 标志与 CMakeLists.txt 中的相同

【参考方案1】：

这是个问题：

set(CUDA_NVCC_FLAGS -gencode arch=compute_20,code=sm_20;-G;-g)

这些标志将导致 nvcc（仅）为 cc 2.0 设备（仅）生成 SASS 代码。这样的 cc2.0 SASS 代码不会在您的 cc5.2 设备 (GTX960) 上运行。 “无效的设备功能”正是在这种情况下尝试启动内核时遇到的错误。由于内核永远不会启动，因此尝试在设备代码中设置断点是行不通的。

我不是 CMake 专家，因此可能还有其他更明智的方法，但尝试解决此问题的一种可能方法是：

set(CUDA_NVCC_FLAGS -gencode arch=compute_52,code=sm_52;-G;-g)

它应该为您的 cc5.2 设备生成代码。毫无疑问，这里还有其他可能的设置，您可能需要阅读this 或nvcc manual 以了解有关针对特定设备的编译选项的更多背景信息。

还要注意-G 生成设备调试代码，如果这是您想要的，这很好。但是，它通常会比没有该开关编译的代码运行得慢。但是，如果要调试，则该开关是必需的。

【讨论】：

谢谢arch=compute_52,code=sm_52 实际上对我有用。

我测试了arch=compute_20,code=\"sm_20,compute_20\"。它也可以正常工作。但为什么？你能对此做进一步的解释吗？

是的，请阅读我链接的答案。该特定组合包括两者 PTX 和 SASS，并且 PTX 可以从 sm_20 到 sm_52 架构进行正向 JIT 编译（在运行时，由驱动程序）以匹配您的 GPU。

再一次，这是一个完美的案例，正确的error checking 可以省去很多麻烦。