0

我在带有 gcc 4.6 的 ubuntu 12.10 上使用 CUDA 5.0 和 GTX 670，并且我编写了一个名为 Grid 的类：

https://github.com/benadler/octocopter/blob/master/basestation/grid.cu

https://github.com/benadler/octocopter/blob/master/basestation/grid.cuh

Grid-class 正在其他两个类中使用：

  • 点云cuda
  • 粒子系统

现在我想在pointcloudcuda和particlesystem的内核中使用Grid的（非静态）方法，即使它们将使用不同的网格（具有不同值的不同网格对象）。因此，对于我使用 Grid 的所有类，我有两个选择：

1）我只是做

Grid hostGrid(...);
cudaMalloc(gridOnDeviceGlobal, sizeof(Grid))
cudaMemcpy(gridOnDeviceGlobal, &hostGrid, sizeof(Grid), cudaMemcpyHostToDevice)
cloudKernel<<< numBlocks, numThreads >>>(someDate, gridOnDeviceGlobal);

这很简单，但是内核必须从全局内存中读取网格值。这可能很慢。

2）由于网格值很少改变，我放了一个

__constant__ Grid myGridForPointCloudCuda

进入pointcloudcuda.cu，连同两个函数

void copyParametersToGpu(Grid *hostGrid)
{
    cudaMemcpyToSymbolAsync(myGridForPointCloudCuda, hostGrid, sizeof(Grid))
}

void getDevicePointerOfGridForPointCloudCuda(Grid** ptr)
{
    cudaGetSymbolAddress((void**)ptr, myGridForPointCloudCuda);
}

现在，在 pointcloudcuda.cpp 中，我可以

Grid hostGrid(...);
copyParametersToGpu(&hostGrid);
Grid* gridOnDeviceConstant;
getDevicePointerOfGridForPointCloudCuda(&gridOnDeviceConstant);
cloudKernel<<< numBlocks, numThreads >>>(someDate, gridOnDeviceConstant);

在我看来，2) 的优点是可以更快地访问内核中的常量内存。但是，在其他地方，我读到这不起作用，因为编译 CUDA 内核的编译器在编译时不知道传递的网格指针是指向全局内存还是常量内存，因此必须使用较慢内存获取指令。

在 Geforce GTX 670 上 2) 会比 1) 快吗？

有没有更好的方法来做我想做的事？我只需要将不同的网格实例传递给内核。在我开始使用多个网格实例之前，一个常量变量是一个舒适且快速的选择。

谢谢！

4

1 回答 1

1

如果您有多个 Grid 实例而不是简单地在常量内存中分配一个 Grid Array，请将 Grid 实例复制到该数组中，并在调用内核时将一个索引传递给 Grid Array，而不是指向特定 Grid 实例的指针。在内核内部使用索引来访问特定的 Grid 实例。

于 2013-02-28T21:13:36.937 回答