减少CPU到GPU数据传输延迟的技术

Question

我一直在探索减少CPU和GPU之间数据传输引起的延迟的方法。当我第一次使用CUDA时，我确实注意到CPU和GPU之间的数据传输需要几秒钟，但我并不在意，因为对于我编写的小程序来说这并不是一个问题。事实上，对于利用GPU的绝大多数程序（包括视频游戏），延迟可能并不是一个大问题，因为它们仍然比在CPU上运行要快得多。

然而，我是一位高性能计算爱好者，当我看到天河一号理论峰值浮点运算速度与实际LINPACK测量性能之间的巨大差距时，我开始担心我的研究方向是否正确。

使用固定内存（page-locked）内存是通过使用cudaHostAlloc()函数进行优化代码以减少延迟的一种方法（非常有效），但我是否还有其他不知道的技术呢？请注意，我谈论的是优化代码，而不是硬件本身（那是NVIDIA和AMD的工作）。

顺便问一下，我知道戴尔和惠普销售Tesla服务器。我很好奇GPU在数据库应用程序中的表现如何，其中您需要从硬盘（HDD或SSD）进行持续读取，这是只有CPU可以执行的操作。

- sj755

---

你可能会觉得这很有趣：http://www.cs.virginia.edu/~skadron/Papers/bakkum_sqlite_gpgpu10.pdf - Mikhail

@Misha 你说得对，这篇文章描述了我所考虑的特定挑战。谢谢... - sj755

关于天河一号低效的问题：尽管GPU拥有大规模算力（这确实经常难以充分利用），但不要忘记GPU内存架构也发挥着重要作用，即使是计算瓶颈核心也可以严重超越CPU实现。 - pszilard

Answer 1

有几种方法可以解决CPU-GPU通讯开销 - 我希望你指的是延迟而不是传输本身的延迟。请注意，我故意使用了“解决”而不是像你所说的“减少”这个词，因为如果您能隐藏它，您不一定需要减少延迟。还要注意的是，我对CUDA更熟悉，因此以下内容仅涉及CUDA，但某些功能也适用于OpenCL。

正如您提到的锁定页面内存（page-locked memory）具有增加性能的目的。此外，人们可以将锁定页面内存映射到GPU上，这种机制使得可以直接访问从GPU内核分配的数据，无需进行额外的数据传输。这种机制称为零拷贝传输，如果数据只读/写一次并伴随着大量计算，且GPU没有单独的内存（移动设备），则非常有用。但是，如果访问零拷贝数据的内核不强烈受计算限制，因此无法隐藏数据访问的延迟，则锁定页面内存但未映射的内存将更有效。另外，如果数据不适合GPU内存，零拷贝仍然可以使用。
请注意，过多的锁定页面内存可能会严重减缓CPU的速度。

从不同角度来解决问题，如tkerwin所提到的，异步传输（相对于CPU线程与GPU进行通信）是通过重叠在CPU上的计算和传输来隐藏CPU-GPU传输延迟的关键。这可以通过cudaMemcpyAsync()以及使用具有异步内核执行的零拷贝来实现。
人们甚至可以使用多个流来重叠传输和内核执行。请注意，流调度可能需要特别注意以实现良好的重叠；Tesla和Quadro显卡具有双DMA引擎，可以同时向GPU传输数据和从GPU传输数据。 此外，CUDA 4.0使得更容易从多个CPU线程中使用GPU，因此在多线程CPU代码中，每个线程都可以将自己的数据发送到GPU并更轻松地启动内核。

最后，GMAC 实现了一个针对 CUDA 的非对称共享内存模型。其中一个非常有趣的特性是提供的一致性模型，尤其是懒惰和滚动更新，使得以阻塞方式仅传输在 CPU 上修改的数据。
更多细节请参见以下论文：Gelado et al. - An Asymmetric Distributed Shared Memory Model for Heterogeneous Parallel Systems。

- pszilard

Answer 2

您可以使用cudaMemcpyAsync()函数将CPU上的工作与内存传输重叠。这不会降低数据传输的延迟，但它可能会提高算法的整体性能。在CUDA C最佳实践指南中有相关信息。

- tkerwin

Answer 3

如果延迟是一个问题，那么值得考虑的是使用AMD融合架构所能带来的权衡。您可以大幅减少延迟，并且在某些情况下比CPU从RAM传输更快。然而，使用精简的非离散GPU会对性能产生影响。

- homemade-jam