Cuda:固定内存零复制问题

Cuda: pinned memory zero copy problems

本文关键字：复制 问题 内存 Cuda      更新时间：2023-10-16

我尝试了此链接中的代码 CUDA 固定内存是零拷贝吗？询问的人声称该程序对他来说效果很好但是在我的身上工作方式不一样如果我在内核中操作它们，这些值不会改变。

基本上我的问题是，我的GPU内存不够，但我想进行需要更多内存的计算。我的程序使用 RAM 内存或主机内存，并且能够使用 CUDA 进行计算。链接中的程序似乎解决了我的问题，但代码没有给出该家伙所示的输出。

任何关于零拷贝内存的帮助或任何工作示例都将很有用。

谢谢

__global__ void testPinnedMemory(double * mem)
{
double currentValue = mem[threadIdx.x];
printf("Thread id: %d, memory content: %fn", threadIdx.x, currentValue);
mem[threadIdx.x] = currentValue+10;
}
void test() 
{
const size_t THREADS = 8;
double * pinnedHostPtr;
cudaHostAlloc((void **)&pinnedHostPtr, THREADS, cudaHostAllocDefault);
//set memory values
for (size_t i = 0; i < THREADS; ++i)
    pinnedHostPtr[i] = i;
//call kernel
dim3 threadsPerBlock(THREADS);
dim3 numBlocks(1);
testPinnedMemory<<< numBlocks, threadsPerBlock>>>(pinnedHostPtr);
//read output
printf("Data after kernel execution: ");
for (int i = 0; i < THREADS; ++i)
    printf("%f ", pinnedHostPtr[i]);    
printf("n");
}

首先，要分配 ZeroCopy 内存，您必须指定cudaHostAllocMapped标志作为cudaHostAlloc的参数。

cudaHostAlloc((void **)&pinnedHostPtr, THREADS * sizeof(double), cudaHostAllocMapped);

pinnedHostPointer仍将仅用于从主机端访问映射的内存。要从设备访问相同的内存，您必须获取指向内存的设备端指针，如下所示：

double* dPtr;
cudaHostGetDevicePointer(&dPtr, pinnedHostPtr, 0);

将此指针作为内核参数传递。

testPinnedMemory<<< numBlocks, threadsPerBlock>>>(dPtr);

此外，您必须将内核执行与主机同步才能读取更新的值。只需在内核调用后添加cudaDeviceSynchronize即可。

链接问题中的代码正常工作，因为提出问题的人正在启用了计算能力 2.0 和 TCC 的 GPU 的 64 位操作系统上运行代码。此配置会自动启用 GPU 的统一虚拟寻址功能，其中设备将主机 + 设备内存视为单个大内存，而不是单独的内存，并且使用 cudaHostAlloc 分配的主机指针可以直接传递到内核。

在您的情况下，最终代码将如下所示：

#include <cstdio>
__global__ void testPinnedMemory(double * mem)
{
    double currentValue = mem[threadIdx.x];
    printf("Thread id: %d, memory content: %fn", threadIdx.x, currentValue);
    mem[threadIdx.x] = currentValue+10;
}
int main() 
{
    const size_t THREADS = 8;
    double * pinnedHostPtr;
    cudaHostAlloc((void **)&pinnedHostPtr, THREADS * sizeof(double), cudaHostAllocMapped);
    //set memory values
    for (size_t i = 0; i < THREADS; ++i)
        pinnedHostPtr[i] = i;
    double* dPtr;
    cudaHostGetDevicePointer(&dPtr, pinnedHostPtr, 0);
    //call kernel
    dim3 threadsPerBlock(THREADS);
    dim3 numBlocks(1);
    testPinnedMemory<<< numBlocks, threadsPerBlock>>>(dPtr);
    cudaDeviceSynchronize();
    //read output
    printf("Data after kernel execution: ");
    for (int i = 0; i < THREADS; ++i)
        printf("%f ", pinnedHostPtr[i]);    
    printf("n");
    return 0;
}
相关文章：