CUDA编程入门：主机端与设备端内存管理

说白了，CUDA 编程里你实际上在跟两个完全独立的内存空间打交道：一个是 CPU 的“主机端内存”，另一个是 GPU 的“设备端内存”。这两个空间不能直接互相访问，数据得靠咱们手动搬运。我个人习惯把这个过程叫“内存搬家三部曲”——分配、拷贝、释放。

为什么非得分这么清楚？

GPU 有自己的显存，CPU 根本访问不到。你想想看，如果直接在 GPU 核函数里用主机端的指针，那肯定报错。我记得刚开始学 CUDA 时，写了个向量加法，传了个主机指针进去，结果内核啥也没干，还静悄悄崩溃了。调试了半天才发现是内存模型没搞对。

所以，你需要记住三件事：

• 分配：在设备端用 cudaMalloc 申请显存
• 拷贝：用 cudaMemcpy 在主机和设备间搬数据
• 释放：用 cudaFree 归还显存

cudaMalloc – 给 GPU 划地盘

函数签名长这样：

cudaError_t cudaMalloc(void **devPtr, size_t size);

第一个参数是二级指针，你要把设备指针的地址传进去。第二个参数是要申请的字节数。返回 cudaError_t，必须检查！我在项目里见过有人直接忽略返回值，结果指针是 NULL，后面 cudaMemcpy 直接崩了。

cudaMemcpy – 数据搬运工的说明书

函数签名：

cudaError_t cudaMemcpy(void *dst, const void *src, size_t count, enum cudaMemcpyKind kind);

这里 kind 有四种常见值，我列个表方便你对照：

方向	枚举值	说明
主机 → 设备	cudaMemcpyHostToDevice	把 CPU 数据拷到 GPU 显存
设备 → 主机	cudaMemcpyDeviceToHost	把 GPU 计算结果取回 CPU
设备 → 设备	cudaMemcpyDeviceToDevice	GPU 内部拷贝（很少手动用）
主机 → 主机	cudaMemcpyHostToHost	调试时偶尔用（几乎不写）

嗯，这里要注意：count 是字节数，不是元素个数。我曾经在项目里把 sizeof(int)*N 写成了 N，结果只拷贝了四分之一的数据，跑出来的结果一塌糊涂。排查半天才发现是长度算错了。

cudaFree – 用完了一定要还

语法很简单：

cudaError_t cudaFree(void *devPtr);

释放掉之前 cudaMalloc 申请的设备指针。千万记得：不要释放同一个指针两次，否则会抛出 cudaErrorInvalidDevicePointer。我自己写框架时，习惯在每次 cudaMalloc 后把指针初始化为 NULL，释放后也置为 NULL，这样能避免误操作。

一个完整的例子：向量加法

看代码比听我说更直观。下面是个简单的向量加法流程：

#include <cuda_runtime.h>
#include <stdio.h>

#define N 1024

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

int main() {
    float *h_a, *h_b, *h_c;        // 主机指针
    float *d_a, *d_b, *d_c;        // 设备指针

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

    // 初始化数据（略）

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

    // 3. 把主机数据拷贝到设备
    cudaMemcpy(d_a, h_a, N * sizeof(float), cudaMemcpyHostToDevice);
    cudaMemcpy(d_b, h_b, N * sizeof(float), cudaMemcpyHostToDevice);

    // 4. 启动内核
    vecAdd<<<(N+255)/256, 256>>>(d_a, d_b, d_c);

    // 5. 把结果拷回主机
    cudaMemcpy(h_c, d_c, N * sizeof(float), cudaMemcpyDeviceToHost);

    // 6. 释放设备内存 & 主机内存
    cudaFree(d_a); cudaFree(d_b); cudaFree(d_c);
    free(h_a); free(h_b); free(h_c);

    return 0;
}

注意看顺序：先分配主机内存，再分配设备内存；然后 H2D 拷贝数据，启动内核，再 D2H 拷贝结果；最后释放所有内存。这个顺序错了，程序就崩了。

常见错误与避坑指南

• 忘记检查返回值：cudaMalloc 和 cudaMemcpy 的返回值必须用 cudaGetLastError() 或直接判断。我曾经在大型训练脚本里漏了检查，结果显存不够时黑盒崩溃，查了一整天才定位到是分配失败。
• 拷贝方向写反：把 H2D 写成 D2H，或者反了。我第一次做时就把结果拷反了，拿到 CPU 的全是垃圾值。
• 释放未分配的指针：比如两次 cudaFree 同一个指针，或者释放了 NULL 指针。嗯，这会导致未定义行为。
• 字节数计算错误：忘了乘 sizeof(type)，或者用了错误的类型大小。我建议用 N * sizeof(*ptr) 这种写法，跟随指针类型自动适应。

下面这张图把整个流程串起来了，你看一眼就能记住内存管理的生命周期：

最后我多说一句：很多新手觉得 cudaFree 可有可无，反正程序退出了系统会回收。但如果你写的是长时间运行的服务（比如训练脚本循环做迭代），忘了释放同一块显存会一直泄漏，直到显存溢出。我自己的习惯是在每个 cudaMalloc 后面写上对应的 cudaFree 注释，写完整个流程再删注释，保证一个都不会漏。