concurrency - 并发CUDA内核执行的优先级

Question

我有两个可以同时执行的内核(A 和 B)。
我需要内核 A 尽快完成(对结果进行 MPI 交换)。所以我可以在一个流中执行它们:A 然后 B。

但是，内核 A 的线程块很少，因此如果我依次运行 A 和 B，则 A 运行时 GPU 不会得到充分利用。

是否可以与具有更高优先级的 A 同时执行 A 和 B？

即，我希望内核 B 中的线程块仅在存在 时才开始执行否 来自内核 A 的非启动块。

据我所知，如果我在一个流中启动内核 A，然后在主机代码的下一行在另一个流中启动内核 B，我不能保证来自 B 的线程块实际上不会首先被执行？

Answer 1

NVIDIA 现在提供了一种对 CUDA 内核进行优先级排序的方法。这是一个相当新的功能，因此您需要升级到 CUDA 5.5 才能使其工作。

对于您的情况，您将启动 kernel A在高优先级 CUDA 流中，然后您启动 kernel B在低优先级的 CUDA 流中。您可能需要的功能是 cudaStreamCreateWithPriority(..., priority) .

要使用此功能，您需要具有 Compute Capability 3.5 或更高版本的 GPU。要检查您的 GPU 是否支持优先级，请查看 cudaDeviceProp::streamPrioritiesSupported .

cudaDeviceGetStreamPriorityRange应该告诉您 GPU 上有多少优先级可用。 cudaDeviceGetStreamPriorityRange 的语法有点不稳定；值得查看 CUDA 手册以了解其工作原理。

更多关于优先级设置的详细文档来自 CUDA Runtime API manual :

cudaError_t cudaStreamCreateWithPriority(cudaStream_t *pStream, 
                                         unsigned int flags, int priority)
Create an asynchronous stream with the specified priority.

Parameters
pStream  = Pointer to new stream identifier 
flags    = Flags for stream creation. See cudaStreamCreateWithFlags for a list of 
           valid flags that can be passed 
priority = Priority of the stream. Lower numbers represent higher priorities. See  
           cudaDeviceGetStreamPriorityRange for more information about the 
           meaningful stream priorities that can be passed.