在for循环中调用CUDA内核

Question

我有一个CUDA核函数，被在for循环内调用。大概是这样的：

for(i=0; i<10; i++) {
   myKernel<<<1000,256>>>(A,i);
}

假设我有一张带有15个流多处理器（SMs）的NVIDIA卡。为简单起见，假设只能将一个块映射到一个SM上，这基本上意味着大部分时间，我会在设备上执行15个块。由于内核执行是异步的，因此i=1的调用将在第一个内核启动后立即排队等待执行（即i=0的那个内核）。

我的问题是：当执行第一个内核（i=0）时，SMs中将只有14个繁忙，然后只有13个，然后只有12个，然后只有11个等等。内核i=1是否会在一个SM可用时立即被发送到设备上执行，还是会等待所有SMs完成处理第一个内核（i=0）？

同时假设我正在一个CUDA流中工作。

- user1586099

Answer 1

在同一流中启动的内核是串行化的。如果有足够的资源（SM、共享内存等），来自不同流的内核调用可能会重叠。

- Eugene

---

对同一流上的操作进行序列化处理。GPU在流完成之前无法读取pushbuffer中的下一个条目。CC2.x-3.0可以支持16个并发内核（如果从不同的流中发出，并且没有错误依赖项或重要状态更改）。GPU将按照通过驱动程序提交内核的顺序将工作分配给SM。这意味着并发工作可以无序完成，但始终会按顺序分配。有关更多信息，请参见http://developer.download.nvidia.com/CUDA/training/StreamsAndConcurrencyWebinar.pdf。 - Greg Smith