我第一次在一台没有图形界面的机器上折腾 CUDA 时，最大的痛点其实不在 kernel 本身，而在于“工程到底怎么被干净地编译出来”。当时随手在 Makefile 里写 nvcc 规则，换一台机器、换一个环境就要重新折腾一遍。后来把整个项目切换到 CMake+CUDA 这套组合，头文件路径、库依赖、GPU 架构、调试符号全部固化在 CMakeLists.txt 里，新机器上一条命令就能重新配置构建。这篇不是官方文档的复述，而是我在实际项目里反复踩坑之后沉淀下来的配置方式，以及那些典型报错的排查路径。

如果你正准备用 CMake 组织 CUDA 代码，或者想把老的 Makefile 工程迁到 CMake，这篇基本可以直接照着抄。代码量不大，但每个关键配置我都会解释为什么这么写，而不是让你无脑复制。

1. 为什么用 CMake 管理 CUDA 项目

1.1 CMake 在 CUDA 构建里做了哪些事

很多第一次接触的人会误以为 CMake 只是把“编译命令”封装了一下，其实 CMake 真正承担的是依赖关系描述和构建系统的生成。

一个 CUDA 工程里有普通 C++ 文件，也有.cu文件。.cu需要由 NVCC 编译，.cpp由宿主 C++ 编译器编译。两者还要共用头文件、链接同一个运行时库。如果手写 Makefile，你需要自己维护每条编译规则、头文件依赖、库路径，换一个平台几乎要重写一遍。CMake 从 3.8 开始原生支持 CUDA 语言，它会在配置阶段主动寻找 CUDA 工具包，识别nvcc的位置，并把.cu文件的编译规则交给各平台的构建系统去执行。

更重要的是，CMake 提供了 target 级别的依赖传递。比如你写target_link_libraries(gpu_ops PRIVATE CUDA::cudart)，那么这个 target 的 include 路径、库路径、编译选项都会跟着目标走。你不需要在每个源文件里手动设置-I、-L、-l。这正好对应了“一处定义、全局生效”的工程化思路。

我自己体感最明显的是跨平台迁移。同样是 CUDA 代码，Linux 上库文件在/usr/local/cuda，Windows 上可能安装在不同的盘符和版本目录，不同环境里的环境变量差异也很大。用 Makefile 硬编码这些路径，迟早会翻车。CMake 会通过find_package(CUDAToolkit)或内置的 CUDA 支持去探测工具包位置，把路径问题交给构建系统自己解决。

1.2 CMake 版本和 CUDA 版本怎么匹配

CMake 对 CUDA 的支持不是一次到位的，不同版本的 CMake，能用的能力差别很大。我整理了一个简单对照表，方便你评估自己的基线。

我现在常用的是 CMake 3.25 以上配合 CUDA 11.8 或 12.x，目前没有遇到明显兼容问题。如果你的系统自带 CMake 版本还是 3.10 上下，第一件事是升级，而不是在旧版本上绕路。因为CUDA_ARCHITECTURES这个变量从 3.18 开始成为主流，晚版本能省掉很多手写-arch的麻烦。

有一个很容易忽略的点：CMake 在配置阶段会检查nvcc和宿主 C++ 编译器是否匹配。比如老版本 CMake 配合特别新的 CUDA 工具包，可能会因为识别不对导致编译失败。所以我倾向于在项目里明确声明 CMake 最低版本，避免队友用旧版本时触发一堆奇怪错误。cmake_minimum_required(VERSION 3.18)不只是一个形式，它确实能拦住一部分兼容性问题。

2. 一个最小可运行的 CMake+CUDA 项目

2.1 三分钟搭一个完整的例子

先从最小的可执行工程开始。目录结构可以这样组织：

demo/ CMakeLists.txt main.cpp kernel.cu

CMakeLists.txt里的内容非常简洁：声明项目、启用 CUDA 语言、指定 GPU 架构，然后添加可执行文件。

cmake_minimum_required(VERSION 3.18) project(cmake_cuda_demo LANGUAGES CXX CUDA) set(CMAKE_CUDA_STANDARD 17) set(CMAKE_CUDA_STANDARD_REQUIRED ON) set(CMAKE_CUDA_ARCHITECTURES "86" "80") add_executable(cmake_cuda_demo main.cpp kernel.cu )

kernel.cu里放一个最简单的向量加法 kernel：

#include <cuda_runtime.h> __global__ void add_kernel(const float* a, const float* b, float* c, int n) { int idx = blockIdx.x * blockDim.x + threadIdx.x; if (idx < n) { c[idx] = a[idx] + b[idx]; } } extern "C" int run_vector_add(const float* a, const float* b, float* c, int n) { add_kernel<<<(n + 255) / 256, 256>>>(a, b, c, n); return cudaGetLastError(); }

main.cpp只负责分配显存、调用 wrapper、把结果拷回来：

#include <cuda_runtime.h> #include <iostream> #include <vector> #include <numeric> int run_vector_add(const float* a, const float* b, float* c, int n); int main() { const int n = 1024; std::vector<float> a(n), b(n), c(n); std::iota(a.begin(), a.end(), 1.0f); std::fill(b.begin(), b.end(), 2.0f); float* da = nullptr; float* db = nullptr; float* dc = nullptr; cudaMalloc(reinterpret_cast<void**>(&da), n * sizeof(float)); cudaMalloc(reinterpret_cast<void**>(&db), n * sizeof(float)); cudaMalloc(reinterpret_cast<void**>(&dc), n * sizeof(float)); cudaMemcpy(da, a.data(), n * sizeof(float), cudaMemcpyHostToDevice); cudaMemcpy(db, b.data(), n * sizeof(float), cudaMemcpyHostToDevice); int err = run_vector_add(da, db, dc, n); if (err != cudaSuccess) return err; cudaDeviceSynchronize(); cudaMemcpy(c.data(), dc, n * sizeof(float), cudaMemcpyDeviceToHost); std::cout << "c[0] = " << c[0] << ", c[512] = " << c[512] << std::endl; cudaFree(da); cudaFree(db); cudaFree(dc); return 0; }

构建只需要两条命令：

cmake -S . -B build -DCMAKE_BUILD_TYPE=Release cmake --build build

生成的可执行文件路径在build/cmake_cuda_demo，运行后能看到c[0]和c[512]的输出。这个例子虽然简单，但已经包含了一个正式 CUDA 工程所需要的全部结构：两种语言混合、kernel 封装、外部声明和链接。

2.2 这些关键 CMake 命令分别干了什么

很多人复制了 CMakeLists.txt 能跑，但一旦要改成自己的项目就不知道怎么动。这里逐个拆开讲一下。

project(cmake_cuda_demo LANGUAGES CXX CUDA)是最核心的一行。它把 CUDA 注册为编译语言。如果你只写了project(xxx LANGUAGES CXX)，把.cu文件放进add_executable，CMake 会直接报错，因为它不知道如何确定.cu文件的链接语言。另一种等价写法是单独调用enable_language(CUDA)。我个人习惯把语言列表直接放到project()里，一眼能看到这个项目用了哪些语言。

set(CMAKE_CUDA_STANDARD 17)和set(CMAKE_CUDA_STANDARD_REQUIRED ON)控制 NVCC 的--std参数。CUDA 代码本质上是 C++ 代码，所以这里起作用的其实是 C++ 标准。如果你用了 C++17 的语法，比如if constexpr、结构化绑定，必须把标准设成 17 或更高。如果不设置CMAKE_CUDA_STANDARD_REQUIRED，CMake 可能会用较低的默认标准，导致编译报语法错误。

set(CMAKE_CUDA_ARCHITECTURES "86" "80")是老版本 CMake 里没有的。它告诉 NVCC 生成哪些 GPU 架构的代码。后面我会单独花一节讲为什么这个变量直接影响你在目标机器上能不能跑起来。

最后是add_executable。它接受多种后缀的源文件，CMake 会根据后缀选择编译器。.cpp走 C++ 编译器，.cu走 NVCC。这也是为什么我不推荐把 CUDA 代码改成.cpp再用-x cu强制 NVCC 编译，后缀本身就是最清晰的调度信号。

3. 编译选项与架构管理的细节

3.1 CUDA_ARCHITECTURES 是容易翻车的点

很多老教程会教你写set(CUDA_NVCC_FLAGS -arch=sm_80)，这是旧时代的做法。从 CMake 3.18 开始，官方推荐使用CMAKE_CUDA_ARCHITECTURES，并且要求显式设置这个变量。如果你没有设置，某些版本会直接报错CUDA_ARCHITECTURES is empty or not set，或者使用了不确定的默认值。

那么"80"和"compute_80"是什么关系？简单说：

• sm_80表示只生成针对某个具体 GPU 架构的 SASS 机器码，优点是可执行文件直接运行，缺点是换个架构就“没有对应的 kernel image”。
• compute_80表示生成 PTX 中间码，运行时再通过驱动 JIT 编译成目标架构 SASS，兼容性好，但第一次运行时有编译开销。

在 CMake 的CUDA_ARCHITECTURES里写"80"，通常会同时生成compute_80和sm_80两种，翻译成 nvcc 参数就是：

-gencode arch=compute_80,code=compute_80 -gencode arch=compute_80,code=sm_80

如果你明确知道只跑在自己的开发机上，可以直接写"80"。如果是发布给别人用，且不清楚对方显卡的 compute capability，一个常见策略是包含较新架构的 SASS，再加上一个较老架构的 PTX。但这样做的代价是编译时间变长、二进制体积变大。不要盲目把从 60 到 90 的架构全写进去，动辄多出几百兆编译产物。

想确认当前显卡的 compute capability，可以在程序里用 CUDA 运行时 API 打印：

int device_id = 0; cudaDeviceProp prop; cudaGetDeviceProperties(&prop, device_id); printf("%d.%d\n", prop.major, prop.minor);

我个人的习惯是在开发阶段使用本机架构，方便快速迭代；发布版本则改成固定的"86" "80"，避免误伤用户机器。如果你把CMAKE_CUDA_ARCHITECTURES放在CMakeLists.txt里写死，换机器时就要手改，这很容易漏掉。所以更好的方案其实是把它交给 CMake preset 或命令行参数，具体我后面会讲。

3.2 按 target 设置编译选项，别污染全局

初学者最容易踩的坑是在 CMakeLists 里写这种代码：

set(CMAKE_CUDA_FLAGS "${CMAKE_CUDA_FLAGS} -use_fast_math")

这种写法对当前目录下的所有 CUDA target 都会生效。如果项目里既有精度要求高的算法，又有性能优先的部分，全局开启-use_fast_math会带来不可控的精度损失。CMake 提供了更细的 target 级编译选项：

target_compile_options(cmake_cuda_demo PRIVATE $<$<COMPILE_LANGUAGE:CUDA>:-lineinfo> )

这里用到了生成器表达式$<COMPILE_LANGUAGE:CUDA>，意思是“只对 CUDA 语言源文件生效”。为什么需要这个判断？因为target_compile_options会把选项同时应用给main.cpp和kernel.cu，而main.cpp用的是 C++ 编译器，它不认识-lineinfo，会直接报错。加了生成器表达式之后，普通 C++ 文件就不会收到这个选项。

常用的几个 NVCC 选项：

• -lineinfo：保留源码行信息，方便后续 profiling 定位到具体 kernel 函数。
• -G：生成设备端调试信息，但会显著降低优化，编译也慢，建议只在 Debug 配置使用。
• -use_fast_math：加快数学运算，但会降低精度，需要业务代码确认可接受。
• --expt-relaxed-constexpr：允许在设备代码里使用较宽松的constexpr规则，部分现代 C++ 库需要它。

举个例子，Debug 下才打开设备调试信息：

target_compile_options(cmake_cuda_demo PRIVATE $<$<AND:$<COMPILE_LANGUAGE:CUDA>,$<CONFIG:Debug>>:-G> )

这样生成器表达式就能根据构建类型动态切换，比手动改全局变量干净很多。

3.3 宿主编译器的匹配问题

NVCC 在编译.cu文件时并不是单独工作，它会把中间代码交给宿主 C++ 编译器处理。所以宿主编译器的版本往往决定了 NVCC 是否正常工作。

如果出现类似unsupported GNU version的报错，通常不是因为你的代码有问题，而是 CUDA 工具包版本不支持当前宿主编译器版本。早期的gcc新版本和 CUDA 的适配经常会有滞后。这时候可以先尝试指定一个更旧、受支持的宿主编译器，或者升级 CUDA 工具包。

CMake 提供变量来指定宿主编译器：

cmake -B build -DCMAKE_CUDA_HOST_COMPILER=/usr/bin/g++-11

这里要注意：CMAKE_CUDA_HOST_COMPILER设置的是 NVCC 调用的宿主编译器，而普通.cpp文件的编译器仍然是CMAKE_CXX_COMPILER。这两者可以不一致，但最好保持一致，否则头文件版本和 ABI 可能产生莫名其妙的差异。

4. 集成 CUDA 库与多语言混合代码

4.1 CUDA 运行时库的链接方式

在一个最简单的 CUDA 可执行文件里，即使你不写find_package(CUDAToolkit)，CMake 也会在启用 CUDA 语言后自动链接一部分运行时库。但当一个工程开始依赖 cuBLAS、cuFFT 等组件时，就需要显式引入 CUDAToolkit 的 target。

推荐的写法是：

find_package(CUDAToolkit REQUIRED) target_link_libraries(cmake_cuda_demo PRIVATE CUDA::cudart CUDA::cublas )

CUDA::cudart是动态运行时库，CUDA::cudart_static是静态运行时库。如果希望最终产物不依赖目标机器上的 CUDA 环境，可以考虑静态链接运行时库。但静态库也不是没有代价，它会增加二进制体积，并且在某些平台上有额外的依赖顺序要求。

CMake 还提供了一个全局开关，可以控制所有 CUDA target 默认使用哪种运行时库模式。需要在project()之前设置才生效：

cmake_minimum_required(VERSION 3.18) set(CMAKE_CUDA_RUNTIME_LIBRARY Static) project(cmake_cuda_demo LANGUAGES CXX CUDA)

我这里要强调一个经验：不要用link_directories()手动指定 CUDA 库路径。一旦你手动指定路径，CMake 的 target 级依赖传递和版本管理就会失效，尤其当系统里有多个 CUDA 版本时，很容易链到错误的库上。正确做法是用find_package(CUDAToolkit)找到的 target 去链接。

4.2 在头文件里同时服务 C++ 和 CUDA

当一个项目同时包含.cpp和.cu，头文件往往会被两个编译器分别处理。最简单的情况是头文件里只放普通 C++ 声明，CUDA kernel 完全封装在.cu内部。但如果一个函数希望在设备代码里也被调用，它就需要标注__device__。问题来了：普通 C++ 编译器根本不认识__device__。

解决方式是用宏隔离：

#pragma once #ifdef __CUDACC__ #define HD __host__ __device__ #else #define HD #endif HD float clamp_01(float v) { return v < 0.0f ? 0.0f : (v > 1.0f ? 1.0f : v); }

当 NVCC 编译.cu文件时，__CUDACC__有定义，所以HD展开为__host__ __device__。当普通 C++ 编译器处理.cpp文件时，HD为空，这个函数就是一个普通内联函数。这种写法在混合代码库里非常实用。

另一个关键点是不要在.cpp文件里直接写 kernel launch 语法，也就是<<<...>>>。这个语法是 NVCC 的扩展，普通 C++ 编译器无法处理。正确做法是把 launch 操作封装到一个.cu文件里，然后通过extern "C"导出普通函数供.cpp调用。前面最小例子里的run_vector_add就是这种模式。

如果你看到明明有 CUDA 代码，编译.cpp时却报错说<<<这里是语法错误，那大概率就是混用了这两种编译方式。

5. 常见问题与排查技巧实录

5.1 一张表先对照常见报错

我在不同机器上折腾 CMake+CUDA 时，遇到的高频问题基本集中在下面几种。先给一张速查表，后面再展开讲。

5.2 编译选项写了却没生效，先看命令

如果你在target_compile_options里写了-lineinfo，运行后 profiler 却看不到行号，先不要怀疑 CMake，而是把构建命令打印出来看看实际传给 NVCC 的参数是什么。

打印方式很简单：

cmake --build build --verbose

如果是 Make 生成器，旧一些的版式也可以：

make VERBOSE=1

看到实际命令后，重点检查两个地方。第一，COMPILE_LANGUAGE:CUDA是否写对了，这个字符串是大小写敏感的，写错会导致生成器表达式恒为 false，选项被直接跳过。第二，CMake 是否重新配置过了。修改 CMakeLists.txt 后，CMake 会根据时间戳自动 rerun，但如果你直接改了缓存变量，有时候并不会完全触发 NVCC 重新编译，稳妥做法是删掉build目录再重新配置。

我实际解决过一次最诡异的 case：命令行里明明有-lineinfo，但编译产物里也没有行号信息。后来发现是因为 build 目录里同时存在多个配置，改版后没有 clean，NVCC 认为产物是最新的就跳过了重编。从那以后，只要修改了架构或者关键编译选项，我都会干脆删掉 build 目录。

5.3 架构写对了还是 no kernel image，怎么办

这个报错一般出现在程序运行时，而不是编译时。它说明最终可执行文件里找不到当前 GPU 能使用的设备代码。常见原因有两种。

第一种是编译架构写得太窄，比如只写了sm_80，但运行机器是sm_86。SASS 通常不会向下兼容到比它更新的架构，所以程序找不到可用的 kernel。这种情况把架构列表改成"80" "86"，或者使用带 PTX 的compute_*就能解决。

第二种是 CMake 缓存里有旧架构。改 CMakeLists.txt 里的CMAKE_CUDA_ARCHITECTURES后，CMake 会重新 configure，但底层的编译产物可能还是旧的。最稳妥的办法是清空 build 目录重新构建。

如果你想确认最终可执行文件里到底有哪些架构，可以用 CUDA 工具包自带的二进制查看工具cuobjdump。执行：

cuobjdump ./build/cmake_cuda_demo

它会列出嵌入的 cubin 和 PTX 信息。只要看到你目标架构对应的条目，基本就能断定编译阶段没问题，问题可能在运行时初始化或驱动版本上。

5.4 找不到 nvcc 或者 CUDA 工具包

CMake 找不到 NVCC 的报错信息一般很直白，比如：

Could not find CUDA compiler.

这种情况通常发生在 CUDA 工具包安装后没有把bin目录放进 PATH。CMake 里有一个专门的变量可以指定 NVCC 路径：

cmake -B build -DCMAKE_CUDA_COMPILER=/usr/local/cuda/bin/nvcc

另一个更工程化的做法是指定CUDAToolkit_ROOT，这样 CMake 会从同一个根目录下找到头文件和库文件：

cmake -B build -DCUDAToolkit_ROOT=/opt/cuda

如果你在容器或 conda 环境里，PATH 和默认目录经常和我们预期的不一样，用这两个变量能快速定位。还有一种比较隐蔽的情况：系统里装了多个 CUDA 版本，CMake 找到了其中一个较旧版本，导致某些新特性报错。这时候用CUDA_HOME环境变量也经常能帮助 CMake 锁定版本，不过要注意CUDA_HOME和CUDAToolkit_ROOT的优先级。最简单的方式还是显式传-DCMAKE_CUDA_COMPILER，一旦编译器确定，工具包其余部分通常会跟着它走。

5.5 Windows 上构建的额外限制

在 Windows 上，最容易出问题的不是 CMake，而是nvcc找不到宿主编译器。

如果你使用 IDE 自带的生成器，CMake 一般会通过 IDE 的配置拿到编译环境，问题不大。但如果你用 Ninja 这个生成器，就要特别注意：必须在已经设置好 C++ 编译器环境变量的终端里执行配置和构建命令。因为 NVCC 在处理.cu文件时，需要调用宿主 C++ 编译器，它依赖INCLUDE、LIB这些环境变量来定位标准库和链接器。直接在普通命令提示符或 PowerShell 里跑 CMake，经常会遇到找不到头文件或无法链接的问题。

如果你在 Windows 上频繁遇到这类问题，我的建议是不要折腾手工环境变量，直接使用 IDE 配套的生成器，或者在工具的“开发人员命令行”入口里重新打开终端再执行 CMake。这个注意事项很容易被新手忽略，因为 Linux 下根本不存在这个概念。

6. 把 CUDA 代码组织成可维护的工程

6.1 把 kernel 拆成独立库

工程一旦变大，不要把几十个.cu文件全部堆在一个可执行文件 target 里。很常见的一种做法是先把 GPU 相关代码抽成静态库，再让可执行文件链接它。

示意结构：

add_library(gpu_ops STATIC src/kernels.cu src/ops_wrapper.cpp ) add_library(gpu_ops::gpu_ops ALIAS gpu_ops) target_include_directories(gpu_ops PUBLIC $<BUILD_INTERFACE:${CMAKE_CURRENT_SOURCE_DIR}/include> ) target_link_libraries(gpu_ops PUBLIC CUDA::cudart) add_executable(app main.cpp ) target_link_libraries(app PRIVATE gpu_ops::gpu_ops)

这样做的好处有三个。第一，改动 kernel 时只需要重新编译gpu_ops这个库，不会把整个可执行文件里所有依赖都拖着重编。第二，如果多个可执行文件都要用到同一批 CUDA kernel，不用重复编译。第三，main.cpp可以不关心 CUDA kernel 的实现细节，它只需要链接gpu_ops::gpu_ops，而链接关系会通过PUBLIC传递得到 CUDA 运行时库。

需要特别注意的是PUBLIC和PRIVATE的选择。如果gpu_ops的公开头文件里需要cuda_runtime.h，那么链接关系必须用PUBLIC，否则下游app虽然链接了库，但编译时找不到头文件。如果只在.cu内部使用 CUDA API，对外头文件不含 CUDA 类型，用PRIVATE就足够了。

6.2 分离编译、缓存和并行提速

CUDA 编译非常吃 CPU。如果你只在命令行里运行cmake --build build，默认情况下可能只用比较少的并行度。这里有两个方向可以提速。

第一个方向是使用 Ninja 生成器，并在构建时显式指定并行任务数：

cmake -S . -B build -G Ninja cmake --build build -j

Ninja 本身比 Makefile 生成的构建系统更快，对增量编译的依赖追踪也更准。因为 CUDA 编译每个文件都很慢，多核并行带来的收益非常明显。

第二个方向是用编译缓存。NVCC 编译重复代码时，缓存工具也可以命中已经编译过的目标文件，避免重复劳动。但缓存对 CUDA 的支持有时候比较微妙，需要确认它已经正确识别了 NVCC 版本和架构参数。如果有多套环境，记得先做一次干净的验证，再引入团队构建流程。

还有一个 CMake 属性值得了解：CUDA_SEPARABLE_COMPILATION。它控制设备代码是否以可分离的方式编译，让不同.cu文件之间的设备函数可以互相调用。默认情况下通常是关闭的，打开后 NVCC 会生成可重定位的设备代码，代价是编译时间更长、生成代码的优化空间可能变小。如果你的项目里没有跨文件调用__device__函数的需求，不要开这个选项。

set_target_properties(gpu_ops PROPERTIES CUDA_SEPARABLE_COMPILATION ON )

6.3 用 CMakePresets 固化你的构建配置

写本文的时候，我已经很少在 CMakeLists.txt 里直接写死CMAKE_CUDA_ARCHITECTURES了。因为同一个项目可能同时要为本机调试和发布环境构建，不同机器上架构不同，写死在 CMakeLists 里就要反复改。CMakePresets 提供了更干净的配置方案。

一个例子：

{ "version": 3, "configurePresets": [ { "name": "release", "generator": "Ninja", "binaryDir": "${sourceDir}/build/release", "cacheVariables": { "CMAKE_BUILD_TYPE": "Release", "CMAKE_CUDA_ARCHITECTURES": "86;80" } } ] }

配置完成后，构建命令只需要：

cmake --preset release cmake --build build/release -j

把一个“项目依赖清单”从 CMakeLists.txt 中剥离出去之后，CMakeLists.txt 更多关注 target 和依赖关系，而构建环境差异交给 preset 管理。这个分层思路对团队协作很有价值。比如有人是 RTX 系列的较新架构，有人是较老的加速卡，各自用不同 preset 就能避免互相改文件。

最后分享一个我自己的习惯：发布给别人的 CUDA 程序，我从来不依赖 CMake 自动检测本机显卡架构，也不会只写一个sm_xx。如果条件允许，我会在目标机器上先查一次 compute capability，然后把准确的架构列表写进 preset。如果没法提前确认，就选择“新一点的 SASS + 较老的 PTX”组合，宁可多花一点编译时间，也比跑到用户现场说“你的机器不支持”好得多。

另外，改完CMAKE_CUDA_ARCHITECTURES之后，如果程序行为没有变化，不要怀疑代码有问题，先删掉 build 目录重新配置一次。这个排查动作成本极低，却经常能解决一堆接口对不上的诡异问题。