1

nvcc 如何处理内核中的 const 指针?

根据 nvidia 的说法,在参数传递期间为指针添加 const 和restrict可以启用 NVCC 进行积极优化,这是否严格遵循 C/C++ 方式?

假设 A 是指向数据缓冲区的指针,该缓冲区可能会被其他线程/流频繁更新,但在此测试内核调用期间内容不会被修改:

test<<<blocks, threads>>>(const int *__restrict__ A, int *__restrict__ B);

那么 NVCC 能否保持这一点的正确性:在每次内核调用时加载 A 中的更新数据,而不是加载一些预先缓存的过时数据?

4

1 回答 1

5

const像 C++ 一样工作。const变量不能更改,编译器会在编译时对其进行检查。编译器仅检查给定范围的 const 正确性,因为可以使用 C 样式转换更改 constness。

restrict以 C 方式工作。当您将指针标记为限制时,编译器假定这些指针没有别名。这是你给定的事实,编译器不会检查这个事实是否正确。

谈到您的问题,NVCC 不会确保内核启动之间全局内存写入和读取的正确性。由于内核启动在 CUDA 中是异步的,因此您必须确保修改这些内存空间的内核不会同时执行。您可以通过同步内存副本和/或cudaDeviceSynchronize(). 如果您同时启动这些内核,则无法确保在来自其他内核的访问之前,不同内核的所有更改都已提交到全局内存。

于 2013-04-16T10:38:23.947 回答