CUDA - 统一内存（至少帕斯卡）答案

【问题标题】：CUDA - Unified memory (Pascal at least)CUDA - 统一内存（至少帕斯卡）
【发布时间】：2018-11-13 17:37:39
【问题描述】：

我想澄清一下统一内存，它是如何工作的以及如何有效地使用它。

据我所知，我们使用cudaMallocManaged(ptr, size); 来分配一个统一内存数组。由于 Pascal 架构，可以将大小设置为大于 GPU 上可用的物理内存。

假设现在我有一个 4GB RAM、32GB RAM 用于主机和一个 1TB 文件的 GC。我想解决这个 1TB 的文件，我该如何处理？

如果我理解的很好，我可以将文件放在统一内存中，但是这个统一数组和文件之间的链接是如何进行的呢？这是否意味着我必须在使用cudaMallocManaged 分配的指针中memcpy 整个文件？

最后，告诉我我是否正确。如果 GPU 出现未命中，CPU 将发送它存储在其 RAM 中的数据，如果不是从磁盘发送。它有点简化，但如果它像这样工作，则意味着数据需要在统一数组中。

感谢您的帮助。

【问题讨论】：

【解决方案1】：

我的回复假设您在 Linux 上运行 CUDA 9.x 或更高版本、Pascal 或 Volta GPU。

您将能够超额订阅 GPU 内存，最高可达主机内存的大小（即主机操作系统允许您分配的任何内存），减去任何内存分配过程中常见的合理数量（您不应该期望分配主机内存的每个最后一个字节，同样不应尝试对托管内存分配做同样的事情。

统一内存与文件或磁盘上存储的任何内容之间没有联系。

正如您可能无法将整个 1TB 文件加载到 32GB RAM 中一样，您也无法使用托管内存一次访问所有文件。主机操作系统允许您分配/加载的量是 GPU 可用的大小。

因此，为了处理该 1TB 文件，您可能需要提出一种算法，将其分解为适合系统 RAM 的片段。这个概念完全独立于托管内存。此后，如果您想使用 CUDA 访问系统 RAM 中的文件，您可以使用托管内存，包括超额订阅，如果您愿意的话。

将文件分解成碎片的确切过程将取决于您正在执行的处理类型，并且对 CUDA 没有特别依赖。

【讨论】：

也许我会问一些愚蠢的问题，但是，你能分配比主机内存更多的内存吗？如果是，会发生什么？操作系统的虚拟化怎么样？不管怎样，你的回答对我帮助很大！
对此没有准确的答案。您可以分配主机操作系统允许您分配的任何内容。它可能超过主机内存。在 32GB 的系统上，它可能不会是 1TB。
过度使用 CPU 内存需要操作系统进行地址转换。据我所知，除非满足特定条件，否则 GPU 无法使用它（请参阅下面的答案）。至于大小，它仅取决于限制和交换空间（或支持文件）。
我不确定我是否同意你的立场。我刚刚在具有 128GB 物理 CPU RAM 的系统上成功调用了 144GB 的cudaMallocManaged。 x86_64、CUDA 9.1、RHEL 7.4、Tesla P100 PCIE。我认为您混淆了在托管分配中过度订阅 CPU 内存的能力，以及在设备代码中使用其他主机指针的能力。我同意后者确实需要 ATS。根据我的测试，托管分配中 CPU 内存的超额订阅将成为操作系统对 CUDA 运行时 API 所说的任何内容的函数，并且在某些情况下，它可以在 x86_64 上运行。
我也尝试过这样做，即过度订阅托管内存。当我使用它时，它最终导致系统完全崩溃，所以我认为它不受支持，正如我在文档中所理解的那样。我可能错了。

【解决方案2】：

在某些系统上可能会超额订阅 CPU 内存。在 NVLink 上使用 Power 9 + V100，您可以使用操作系统的地址转换服务 (ATS)，表示为 here。

这样做，可以使用来自 GPU 的 1TB 数据，即使主机系统上的 RAM 量要少得多。要做的操作如下：

操作系统的分页机制会按需调入调出文件的块，而 GPU 将依赖 ATS 进行此操作。

没有提到在 x86_64 和/或上一代 GPU 和/或 PCI-Express 连接系统上进行此类练习，也没有成功测试。

【讨论】：

请注意，即使没有超额订阅，使用“传递给内核”的主机指针也需要 ATS（当前）。而且我相信在 x86_64 上可能会超额订阅。但我同意在设备代码中使用主机指针（即不是在 CUDA 运行时创建或注册的任何东西）需要 ATS，有或没有超额订阅，并且 ATS（或类似的东西）目前在任何 x86_64 平台上都不可用，我知道。