如何以编程方式确定持久内核的正确启动参数？

Question

以编程方式确定持久内核启动参数的正确方法是什么？我发现的所有示例都使用硬编码值。

以下是正确的吗？

cudaDeviceProp props;

cudaGetDeviceProperties(&props, 0);

int blockCount = props.maxBlocksPerMultiProcessor * props.multiProcessorCount;
int blockThreadCount = props.maxThreadsPerMultiProcessor / props.maxBlocksPerMultiProcessor;

//  Gives <<<1312, 96>>> on a RTX 3090
PersistentKernel<<<blockCount, blockThreadCount>>>(...);

Answer 1

Is the following correct?

没有

使用cudaOccupancyMaxPotentialBlockSize。这将为您提供当前设备的网格大小和块大小，从而以最少的块数最大化给定内核的占用率。这是给定持久内核的最佳启动参数。

请注意，返回的块和网格维度是标量。您可以自由地将它们重塑为多维 dim3 块 and/or 网格维度，这些维度保留每个块的线程总数以及由 API.

返回的块