首页 文章

Cuda：内核启动队列

提问于
浏览 次
3

我没有找到关于内核启动操作机制的更多信息 . API说要看CudaProgGuide . 我也找不到那么多 .
由于内核执行是异步的，并且一些机器支持并发执行，因此我认为内核有一个队列 .

Host code:      
    1. malloc(hostArry, ......);  
    2. cudaMalloc(deviceArry, .....);  
    3. cudaMemcpy(deviceArry, hostArry, ... hostToDevice);
    4. kernelA<<<1,300>>>(int, int);  
    5. kernelB<<<10,2>>>(float, int));  
    6. cudaMemcpy(hostArry, deviceArry, ... deviceToHost);  
    7. cudaFree(deviceArry);

第3行是同步的 . 第4行和第5行是异步的，并且机器支持并发执行 . 所以在某些时候，这两个内核都在GPU上运行 . （在kernelA完成之前，kernelB有可能启动并完成 . ）当发生这种情况时，主机正在执行第6行 . 第6行与复制操作是同步的，但没有什么能阻止它在kernelA之前执行或者kernelB已经完成了 .

1）GPU中是否有内核队列？ （GPU是否阻止/停止主机？）
2）主机如何知道内核已经完成，并且将设备的结果发送到主机是"safe"？

cuda

2 回答

  • 3

    是的，GPU上有各种队列，驱动程序管理这些队列 .

    异步调用或多或少立即返回 . 在操作完成之前，同步调用不会返回 . 内核调用是异步的 . 如果它们是异步的，则大多数其他CUDA运行时API调用由后缀 Async 指定 . 所以回答你的问题：

    1）GPU中是否有内核队列？ （GPU是否阻止/停止主机？）

    有各种队列 . GPU在同步调用中阻塞/停止主机，但内核启动不是同步操作 . 它在内核完成之前立即返回，也许在内核启动之前返回 . 将操作启动到单个流中时，该流中的所有CUDA操作都将被序列化 . 因此，即使内核启动是异步的，您也不会发现启动到同一个流的两个内核的重叠执行，因为CUDA子系统保证流中的给定CUDA操作在相同流中的所有先前CUDA操作之前都不会启动完了 . 对于空流（如果您未在代码中明确地调用流，则使用的流）还有其他特定规则，但前面的描述足以理解此问题 .

    2）主机如何知道内核已经完成，将设备的结果发送到主机是否“安全”？

    由于将结果从设备传输到主机的操作是 CUDA 调用（cudaMemcpy ...），并且它在与前面操作相同的流中发出，因此设备和CUDA驱动程序管理cuda调用的执行顺序，以便在发出到同一个流的所有先前CUDA调用完成之前，cudaMemcpy才会开始 . 因此，即使您使用 cudaMemcpyAsync ，在内核调用完成之后，在相同流中的内核调用之后发出的 cudaMemcpy 保证不会启动 .

    回复于
  • 0

    您可以在内核调用后使用cudaDeviceSynchronize()来保证为设备请求的所有先前任务都已完成 . 如果kernelB的结果与kernelA上的结果无关，则可以在内存复制操作之前设置此函数 . 如果没有，则需要在调用kernelB之前阻塞设备，从而导致两个阻塞操作 .

    回复于

相关问题