我第一次在一台没有图形界面的机器上折腾 CUDA 时最大的痛点其实不在 kernel 本身而在于“工程到底怎么被干净地编译出来”。当时随手在 Makefile 里写 nvcc 规则换一台机器、换一个环境就要重新折腾一遍。后来把整个项目切换到 CMakeCUDA 这套组合头文件路径、库依赖、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/cudaWindows 上可能安装在不同的盘符和版本目录不同环境里的环境变量差异也很大。用 Makefile 硬编码这些路径迟早会翻车。CMake 会通过find_package(CUDAToolkit)或内置的 CUDA 支持去探测工具包位置把路径问题交给构建系统自己解决。1.2 CMake 版本和 CUDA 版本怎么匹配CMake 对 CUDA 的支持不是一次到位的不同版本的 CMake能用的能力差别很大。我整理了一个简单对照表方便你评估自己的基线。CMake 版本对 CUDA 的重要支持实际使用建议3.8首次支持 CUDA 作为 C 编译语言太老不推荐3.12支持CUDA_STANDARD属性提供CUDA::cudart_static等 target可以考虑3.18引入CUDA_ARCHITECTURES替代老的CUDA_ARCH_BIN强烈建议使用3.24对架构列表和编译选项的处理更完善推荐使用我现在常用的是 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. 一个最小可运行的 CMakeCUDA 项目2.1 三分钟搭一个完整的例子先从最小的可执行工程开始。目录结构可以这样组织demo/ CMakeLists.txt main.cpp kernel.cuCMakeLists.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::vectorfloat 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_castvoid**(da), n * sizeof(float)); cudaMalloc(reinterpret_castvoid**(db), n * sizeof(float)); cudaMalloc(reinterpret_castvoid**(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_TYPERelease 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_executableCMake 会直接报错因为它不知道如何确定.cu文件的链接语言。另一种等价写法是单独调用enable_language(CUDA)。我个人习惯把语言列表直接放到project()里一眼能看到这个项目用了哪些语言。set(CMAKE_CUDA_STANDARD 17)和set(CMAKE_CUDA_STANDARD_REQUIRED ON)控制 NVCC 的--std参数。CUDA 代码本质上是 C 代码所以这里起作用的其实是 C 标准。如果你用了 C17 的语法比如if constexpr、结构化绑定必须把标准设成 17 或更高。如果不设置CMAKE_CUDA_STANDARD_REQUIREDCMake 可能会用较低的默认标准导致编译报语法错误。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 -archsm_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 archcompute_80,codecompute_80 -gencode archcompute_80,codesm_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 一张表先对照常见报错我在不同机器上折腾 CMakeCUDA 时遇到的高频问题基本集中在下面几种。先给一张速查表后面再展开讲。报错现象可能原因处理方向CUDA_ARCHITECTURES is empty or not set没有设置 GPU 架构列表在 CMakeLists 或命令行设置CMAKE_CUDA_ARCHITECTURESfatal error: cuda_runtime.h: No such file or directory没有找到 CUDA 工具包或 include 路径未配置设置CUDAToolkit_ROOT确认 NVCC 存在nvcc fatal : Unsupported gpu architecture写了当前 CUDA 版本不支持的架构降低架构列表或升级 CUDA 工具包/usr/bin/ld: cannot find -lcudartCUDAToolkit 库路径不对使用find_package(CUDAToolkit)的 target 链接运行时no kernel image is available编译的架构和实际显卡不匹配扩展CMAKE_CUDA_ARCHITECTURES支持目标架构Windows 下nvcc找不到宿主编译器没有在正确环境 shell 中执行 CMake使用 IDE 生成器或在工具链命令行里构建5.2 编译选项写了却没生效先看命令如果你在target_compile_options里写了-lineinfo运行后 profiler 却看不到行号先不要怀疑 CMake而是把构建命令打印出来看看实际传给 NVCC 的参数是什么。打印方式很简单cmake --build build --verbose如果是 Make 生成器旧一些的版式也可以make VERBOSE1看到实际命令后重点检查两个地方。第一COMPILE_LANGUAGE:CUDA是否写对了这个字符串是大小写敏感的写错会导致生成器表达式恒为 false选项被直接跳过。第二CMake 是否重新配置过了。修改 CMakeLists.txt 后CMake 会根据时间戳自动 rerun但如果你直接改了缓存变量有时候并不会完全触发 NVCC 重新编译稳妥做法是删掉build目录再重新配置。我实际解决过一次最诡异的 case命令行里明明有-lineinfo但编译产物里也没有行号信息。后来发现是因为 build 目录里同时存在多个配置改版后没有 cleanNVCC 认为产物是最新的就跳过了重编。从那以后只要修改了架构或者关键编译选项我都会干脆删掉 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 -jNinja 本身比 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 目录重新配置一次。这个排查动作成本极低却经常能解决一堆接口对不上的诡异问题。