从零搭起CUDA开发环境：我的安装避坑与第一个向量加法

说实话，很多同学学CUDA卡在第一步——环境没配好，编译都过不去。我个人折腾过Linux和Windows两套环境，踩过不少坑。这一章咱们就把CUDA安装、nvcc编译器和第一个程序一口气打通关。

CUDA安装与环境配置：讲点实在的

先说你最关心的问题：到底该装哪个版本？我的建议是，如果你的显卡不太老，直接上CUDA 12.x。但如果你在维护老项目，可能还得用11.x。

下载地址很简单：NVIDIA官网Developer专区。别去百度搜第三方链接，我吃过那个亏——下了个改版包，编译怎么都报错。

安装过程我不啰嗦了，注意两点：
第一，Windows下一定要勾选「Visual Studio Integration」，否则vs里找不到cu文件模板。
第二，安装完后务必检查环境变量。Linux下加这几行到 ~/.bashrc：

export PATH=/usr/local/cuda-12/bin:$PATH
export LD_LIBRARY_PATH=/usr/local/cuda-12/lib64:$LD_LIBRARY_PATH
export CUDA_HOME=/usr/local/cuda-12

验证安装最简单的方法：

nvcc --version

如果能看到版本号，恭喜你，环境搭好了。我曾经在一个服务器上配完看不到版本，后来发现装的是cuda-driver包，不是完整的toolkit——这种坑很恶心。

nvcc编译器：你和GPU沟通的桥梁

nvcc是什么？说白了就是NVIDIA的C++编译器，但它认识.cu文件里那些奇怪的<<<>>>语法。你写的核函数、设备代码，都是通过nvcc编译成GPU能执行的PTX或二进制。

它的工作流程分两步：
先编译设备代码（也就是核函数），生成 .fatbin 或 .cubin。
然后主机代码用常规C++编译器（gcc/clang/MSVC）编译，中间通过cudart链接。

我个人习惯用nvcc直接编译整个项目，简单粗暴：

nvcc -o vector_add vector_add.cu

如果项目大了，也可以用CMake管理，加一行：

enable_language(CUDA)

但初期我建议你就用单命令行，跑通了再上构建系统。

第一个CUDA程序：向量加法

好了，咱们写个经典向量加法。我带你一行行拆解，以后你写任何CUDA程序，骨架都差不多。

#include <stdio.h>

// GPU核函数：每个线程负责一个元素
__global__ void vecAdd(const float *a, const float *b, float *c, int n) {
    int i = blockIdx.x * blockDim.x + threadIdx.x;
    if (i < n) {
        c[i] = a[i] + b[i];
    }
}

int main() {
    int n = 1 << 20; // 1048576个元素
    size_t bytes = n * sizeof(float);

    // 1. 分配主机内存
    float *h_a = (float*)malloc(bytes);
    float *h_b = (float*)malloc(bytes);
    float *h_c = (float*)malloc(bytes);

    // 初始化数据
    for (int i = 0; i < n; i++) {
        h_a[i] = 1.0f;
        h_b[i] = 2.0f;
    }

    // 2. 分配设备内存
    float *d_a, *d_b, *d_c;
    cudaMalloc(&d_a, bytes);
    cudaMalloc(&d_b, bytes);
    cudaMalloc(&d_c, bytes);

    // 3. 从主机拷贝到设备
    cudaMemcpy(d_a, h_a, bytes, cudaMemcpyHostToDevice);
    cudaMemcpy(d_b, h_b, bytes, cudaMemcpyHostToDevice);

    // 4. 启动核函数
    int blockSize = 256;
    int gridSize = (n + blockSize - 1) / blockSize;
    vecAdd<<<gridSize, blockSize>>>(d_a, d_b, d_c, n);

    // 5. 拷贝结果回主机
    cudaMemcpy(h_c, d_c, bytes, cudaMemcpyDeviceToHost);

    // 6. 验证
    for (int i = 0; i < 10; i++) {
        printf("%f + %f = %f\n", h_a[i], h_b[i], h_c[i]);
    }

    // 7. 清理
    cudaFree(d_a);
    cudaFree(d_b);
    cudaFree(d_c);
    free(h_a);
    free(h_b);
    free(h_c);

    return 0;
}

你看，总共就7步。我在项目里每次写新的核函数都照着这个模板复制粘贴，省脑子。编译运行：

nvcc -o vector_add vector_add.cu
./vector_add

如果输出正确，说明你的第一个CUDA程序跑通了。

主机与设备代码分离：第一个编程模型分界线

你可能注意到了，上面的代码里，__global__ 修饰的是设备代码（跑在GPU上），main() 是主机代码（跑在CPU上）。这就是CUDA最基本的分离思想。

实际工程中，你不会把核函数和主调代码混在一个文件里。我一般这么组织：

• vec_add_kernel.cu — 只放核函数定义，以及封装核函数调用的包装函数
• main.cpp — 纯C++代码，包含内存管理、数据准备、调用包装函数
• cuda_utils.h — 声明包装函数，以及错误检查宏

为什么这么拆？因为nvcc和普通C++编译器各管各的。.cu文件用nvcc编译，.cpp文件用gcc/clang编译，最后链接。拆开之后，你的代码更干净，编译也更清晰。

举个例子，你可以把核函数放在一个单独的文件：

// vec_add_kernel.cu
#include "cuda_utils.h"

__global__ void vecAddKernel(const float *a, const float *b, float *c, int n) {
    int i = blockIdx.x * blockDim.x + threadIdx.x;
    if (i < n) c[i] = a[i] + b[i];
}

void vecAdd(const float *a, const float *b, float *c, int n) {
    float *d_a, *d_b, *d_c;
    size_t bytes = n * sizeof(float);
    cudaMalloc(&d_a, bytes);
    cudaMalloc(&d_b, bytes);
    cudaMalloc(&d_c, bytes);
    cudaMemcpy(d_a, a, bytes, cudaMemcpyHostToDevice);
    cudaMemcpy(d_b, b, bytes, cudaMemcpyHostToDevice);

    int blockSize = 256;
    int gridSize = (n + blockSize - 1) / blockSize;
    vecAddKernel<<<gridSize, blockSize>>>(d_a, d_b, d_c, n);

    cudaMemcpy(c, d_c, bytes, cudaMemcpyDeviceToHost);
    cudaFree(d_a);
    cudaFree(d_b);
    cudaFree(d_c);
}

然后主函数只需要调用包装函数：

// main.cpp
#include "cuda_utils.h"
int main() {
    // ... 分配主机内存 ...
    vecAdd(h_a, h_b, h_c, n);
    // ... 验证 ...
    return 0;
}

编译命令也变了：

nvcc -c vec_add_kernel.cu -o vec_add_kernel.o
g++ -c main.cpp -o main.o
g++ main.o vec_add_kernel.o -L/usr/local/cuda/lib64 -lcudart -o vector_add

嗯，这里要注意，链接时要加上 -lcudart，否则报一堆未定义引用。

一张图看懂本章核心

下面这张SVG把整个流程画清楚了：从主机内存到设备内存，核函数执行，再拷回来。你看懂这张图，本章你就掌握了。

好了，环境搭好，第一个程序跑通，主机设备分离也清楚了。接下来你就可以往里面塞你自己的算法了——别忘了先调 cudaMalloc 再传数据，这个顺序不能乱。