CUDA 统一内存：让 GPU 与 CPU 协作无间的魔法-HQY 一个和谐有爱的空间

25

2025
03
20:30:58

CUDA 统一内存：让 GPU 与 CPU 协作无间的魔法

CUDA 的统一内存（Unified Memory）是通过 cudaMallocManaged 函数和 __managed__ 关键字实现的主机与设备的透明化内存访问。其核心原理是将物理存储位置抽象为统一的虚拟地址空间，当 CPU 或 GPU 访问数据时，系统自动完成数据迁移（按需分页迁移），开发者无需手动调用 cudaMemcpy。这对于简化代码非常有用，特别是对于初学者或者需要快速原型开发的情况。

`cudaMallocManaged` 简介

cudaMallocManaged 是 CUDA 提供的一个函数，用于分配统一内存。统一内存是一种虚拟内存空间，它可以被 CPU 和 GPU 共同访问，CUDA 运行时系统会自动处理内存的迁移和同步，使得开发者无需手动在 CPU 和 GPU 之间复制数据。

#include <cstdio>
#include <cuda_runtime.h>

__global__ void addKernel(float* A, 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;
    float *A, *B, *C;

    // 分配统一内存
    cudaMallocManaged(&A, N*sizeof(float));
    cudaMallocManaged(&B, N*sizeof(float));
    cudaMallocManaged(&C, N*sizeof(float));

    // 主机初始化
    for (int i = 0; i < N; i++) {
        A[i] = 1.0f;
        B[i] = 2.0f;
    }

    // 启动内核
    addKernel<<<(N+255)/256, 256>>>(A, B, C, N);

    // 等待 GPU 完成
    cudaDeviceSynchronize();

    // 验证结果
    printf("C[0] = %f\n", C[0]);

    // 释放内存
    cudaFree(A);
    cudaFree(B);
    cudaFree(C);
    return0;
}

为了快速验证，可以将上述代码保存为 test.cu，然后使用 nvcc 进行编译，编译和执行指令如下：

nvcc -o test test.cu && ./test

managed 简介

__managed__ 是 CUDA 提供的一个关键字，用于声明统一内存变量。使用 __managed__ 关键字声明的变量会被自动分配到统一内存中，同样可以被 CPU 和 GPU 共同访问。需要说明的是 __managed__ 变量需要声明为全局变量，因此变量生命周期与应用程序相同。

#include <iostream>
#include <cuda_runtime.h>

__managed__ int data[1024]; // 自动分配到统一内存中，可以被 CPU 和 GPU 共同访问

__global__ void kernel(int* data, int n) {
    int idx = threadIdx.x + blockIdx.x * blockDim.x;
    if (idx < n) {
        data[idx] *= 2;
    }
}

int main() {
    int n = 1024;

    // 初始化数据
    for (int i = 0; i < n; i++) {
        data[i] = i;
    }

    // 定义线程块和网格大小
    int blockSize = 256;
    int gridSize = (n + blockSize - 1) / blockSize;

    // 调用核函数
    kernel<<<gridSize, blockSize>>>(data, n);

    // 同步设备
    cudaDeviceSynchronize();

    // 输出结果
    for (int i = 0; i < 10; i++) {
        std::cout << data[i] << " ";
    }
    std::cout << std::endl;

    return0;
}

代码执行方式同上一个例子。

需要注意的是，__managed__ 在使用时需要注意以下事项：

• __managed__ 变量的地址不是常量表达式。
常量表达式是在编译时就能计算出结果的表达式。常量表达式通常用于数组大小的指定、枚举值的初始化、模板参数等需要在编译时确定值的地方。而在 CUDA 中，使用 __managed__ 修饰符声明的变量，其地址在编译时是无法确定的，因为托管（managed）内存地址的分配是在运行时由 CUDA 运行时系统动态处理的。这意味着不能将 __managed__ 修饰符声明的变量的地址用作常量表达式。
• __managed__ 变量不能使用 const 限定符修饰。
• __managed__ 变量不能是引用类型。
• 当 CUDA 运行时可能处于无效状态时，不得使用托管变量的地址或值，这包括在具有静态或线程局部存储期的对象的静态/动态初始化或销毁期间。
• 当包含托管变量的 CUDA 程序在具有多个 GPU 的执行平台上运行时，变量仅分配一次，而不是每个 GPU 分配一次。
• 在主机或设备上执行的函数不可以使用没有外部链接性的托管变量。

其实没必要过分担心这些约束条件，因为 __managed__ 变量的使用目的就是为了简化开发流程和快速原型开发，也就是说通常的使用场景不会太复杂，毕竟要想实现高性能的程序，往往需要精细化控制代码的执行，通常会使用额外的方法来实现。

统一内存的特性

• 自动内存迁移：
统一内存由 CUDA 运行时自动管理，数据在 CPU 或 GPU 访问时按需迁移，开发者无需手动调用 cudaMemcpy。
• 单一指针访问：
通过 cudaMallocManaged 分配的内存，主机和设备代码使用同一个指针，简化编程模型。
• 支持并发访问：
在支持计算能力 6.0+（如 Pascal 架构及更高）的 GPU 上，CPU 和 GPU 可以同时访问统一内存（需注意同步）。

统一内存的使用场景

• 简化代码：
适合快速原型开发，避免手动管理 cudaMemcpy。
• 复杂内存访问模式：
当数据被 CPU 和 GPU 交替访问且模式难以预测时（如图遍历、动态数据结构）。
• 显存不足的辅助：
允许部分数据驻留在主机内存，扩展可用内存容量（但可能增加延迟）。

在 CUDA 编程时，统一内存分配和管理虽然为我们带来了便利，但频繁的数据迁移会导致性能下降，因此可以通过一些技巧来提升性能。

性能优化技巧

手动预取数据

// 在 GPU 访问前预取数据到显存
cudaMemPrefetchAsync(data, size, deviceId);

// 在 CPU 访问前预取回主机内存
cudaMemPrefetchAsync(data, size, cudaCpuDeviceId);

内存访问提示

// 提示数据主要被 GPU 访问
cudaMemAdvise(data, size, cudaMemAdviseSetPreferredLocation, deviceId);

// 提示数据将被频繁读取
cudaMemAdvise(data, size, cudaMemAdviseSetReadMostly, deviceId);

避免乒乓访问

• 尽量让数据在 GPU 或 CPU 上集中访问，减少来回迁移。
• 在支持 NVLink 的系统中，统一内存性能更优（高带宽、低延迟）。

其他优化建议

• 使用 cudaStreamAttachMemAsync 将内存绑定到特定 CUDA 流，提升并发性。
• Pascal 架构（SM6.0+）支持原子操作和更细粒度迁移，Kepler/Maxwell 需注意访问冲突。

总结

cudaMallocManaged 和 __managed__ 通过统一内存模型简化了 CUDA 编程，这非常适合 CUDA 初学者以及需要快速原型开发的场景，但统一内存虽然用法简单，却对开发者隐藏了一些控制细节，因而失去了一些控制灵活性和性能的降低，需结合具体场景权衡便利性与性能。合理使用预取、内存建议和同步机制，可最大化其优势。

推荐本站淘宝优惠价购买喜欢的宝贝: