CUDA中GPU全局内存分配方法及host端的访问方式

OREH_HERO

已于 2023-11-05 21:12:56 修改

阅读量851

点赞数

文章标签：算法

于 2023-11-05 16:42:14 首次发布

本文链接：https://blog.csdn.net/OREH_HERO/article/details/134231721

版权

本文介绍了CUDA中全局内存的不同分配方式（静态、动态和CUDARuntimeAPI），以及host如何访问GPU全局内存的规则，通过示例展示了内存复制操作和核函数的使用。

摘要生成于 C知道，由 DeepSeek-R1 满血版支持，前往体验 >

1.GPU全局内存分配

CUDA中，全局内存（global memory）可以用静态方法（使用__device__）、动态方法（使用设备的malloc或new）和CUDA runtime API（使用cudaMalloc）分配。

2.host端访问GPU全局内存

需要注意的是，这三种方法分配的内存在device中的访问方式相同，但它们与host有不同的访问方式。

host访问静态分配的内存时，需要使用cudaMemcpyToSymbol和cudaMemcpyFromSymbol;

访问CUDA runtime API分配的全局内存需要使用cudaMolloc和cudaMemcpy类型的函数；

而访问直接使用设备malloc或new来动态分配的全局内存不能直接从host端访问。

代码案例如下，后续补充说明。

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

// 静态分配的全局内存
__device__ int globalData[256]; 

// 核函数
template<typename T>
__global__ void memoryExample(T* deviceData)
{
    int tid = threadIdx.x + blockIdx.x * blockDim.x;

    // 全局内存读写
    globalData[tid] = deviceData[tid];   
}

int main(){
    const int dataSize = 256;
    int hostData[dataSize];
    int hostResult[dataSize];

    for (int i = 0; i < dataSize; ++i)
        hostData[i] = i;

    // GPU内存声明与拷贝
    int* deviceData = nullptr;
    cudaMalloc((void**)&deviceData, dataSize * sizeof(int));
    cudaMemcpy(deviceData, hostData, dataSize * sizeof(int), cudaMemcpyHostToDevice);

    // 调用核函数
    memoryExample << <1, dataSize >> > (deviceData);

    // 静态分配的全局函数拷贝与输出
    cudaMemcpyFromSymbol(hostResult, globalData, dataSize * sizeof(int));

    for (int i = 0; i < dataSize; ++i)
        printf("%d ", hostResult[i]);
    printf("\n");

    cudaFree(deviceData);
    return 0;
}