可以直接从gpu访问硬盘吗？答案

【问题标题】：Is it possible to access hard disk directly from gpu?可以直接从gpu访问硬盘吗？
【发布时间】：2015-02-01 15:20:00
【问题描述】：

是否可以直接从 GPU (CUDA/openCL) 访问硬盘/闪存盘并直接从 GPU 的内存加载/存储内容？

我试图避免将内容从磁盘复制到内存，然后再将其复制到 GPU 的内存。

我阅读了有关 Nvidia GPUDirect 的信息，但不确定它是否符合我上面的解释。它讨论了远程 GPU 内存和磁盘，但在我的例子中，磁盘是 GPU 本地的。

基本思想是加载内容（类似于 dma）-> 执行一些操作 -> 将内容存储回磁盘（再次以 dma 方式）。

我试图在这里尽可能少地涉及 CPU 和 RAM。

请随时提供有关设计的任何建议。

【问题讨论】：

有人能解释一下这个问题有什么问题吗？
没有主机干预是不可能的。主机拥有磁盘驱动器。 GPUDirect 从根本上是用于在 PCIE 设备之间传输数据。如果您有自己的 PCIE HDD 控制器，在与 GPU 相同的 PCIE 结构上，并且可以访问设备驱动程序源代码，您可以想象编写一个 GPUDirect RDMA 驱动程序，该驱动程序允许从 GPU 直接传输到磁盘。（它仍然需要主机干预才能设置。）实际上，没有人认为这是您想要承担的工作量。
实际上，系统内存吞吐量（25-50 GB/秒）和 PCIe gen3 吞吐量（10-12 GB/秒）与 SSD 吞吐量（0.5 GB/秒）相比非常高，以至于在通过主机移动数据时，对 GPU 磁盘传输的吞吐量的影响应该很小。延迟可能是另一回事，但问题并未说明具体的延迟或吞吐量要求。
@Siddharth 你可能想看看this GTC 2014 presentation，它讨论了 GPUdirect RDMA 访问类似 SSD 的存储。
@bit2shift 据我所知，Radeon SSG 产品在这一点上是蒸汽软件，或者，如果你愿意的话，是概念验证。据我所知，它将两个 SSD（在 RAID0 配置中）耦合到 GPU，每个都有一个 PCIe gen3 x4 链接，这意味着 SSD 到 GPU 链接的总吞吐量是 GPU 的普通 PCIe x16 链接的一半到系统。它可能以较低的延迟这样做，但在我看来并不像游戏规则改变者。

标签： cuda parallel-processing opencl gpu

【解决方案1】：

对于寻找这个的其他人来说，“懒惰的取消固定”或多或少地做了我想要的。

查看以下内容是否对您有帮助。

为 GPUDirect 使用 RDMA 的最直接的实现是每次传输前固定内存并在传输后立即取消固定做完了。不幸的是，这通常会表现不佳，因为固定和取消固定内存是昂贵的操作。剩下的但是，可以执行执行 RDMA 传输所需的步骤快速不进入内核（DMA列表可以被缓存和使用 MMIO 寄存器/命令列表重放）。

因此，延迟取消固定内存是高性能 RDMA 的关键执行。它的含义是保持内存固定不变传输完成后。这利用了以下事实：未来的 DMA 可能会使用相同的内存区域传输因此延迟取消固定节省了固定/取消固定操作。

延迟取消固定的示例实现将保持一组固定内存区域，并且只取消固定其中的一些（例如，最少最近使用的一个）如果区域的总大小达到了一些阈值，或者如果由于 BAR 空间而固定新区域失败耗尽（参见 PCI BAR 大小）。

这是指向application guide 和nvidia docs 的链接。

【讨论】：

【解决方案2】：

为了使用这个特性，我在 Windows x64 上写了一个小例子来实现它。在这个例子中，内核“直接”访问磁盘空间。实际上，正如@RobertCrovella 之前提到的，操作系统正在完成这项工作，可能需要一些 CPU 工作；但没有补充编码。

__global__ void kernel(int4* ptr)
{
    int4 val ; val.x = threadIdx.x ; val.y = blockDim.x ; val.z = blockIdx.x ; val.w = gridDim.x ;
    ptr[threadIdx.x + blockDim.x * blockIdx.x] = val ;
    ptr[160*1024*1024 + threadIdx.x + blockDim.x * blockIdx.x] = val ;
}

#include "Windows.h"

int main()
{
    // 4GB - larger than installed GPU memory
    size_t size = 256 * 1024 * 1024 * sizeof(int4) ;

    HANDLE hFile = ::CreateFile ("GPU.dump", (GENERIC_READ | GENERIC_WRITE), 0, 0, CREATE_ALWAYS, FILE_ATTRIBUTE_NORMAL, NULL) ;

    HANDLE hFileMapping = ::CreateFileMapping (hFile, 0, PAGE_READWRITE, (size >> 32), (int)size, 0) ;

    void* ptr = ::MapViewOfFile (hFileMapping, FILE_MAP_ALL_ACCESS, 0, 0, size) ;

    ::cudaSetDeviceFlags (cudaDeviceMapHost) ;

    cudaError_t er = ::cudaHostRegister (ptr, size, cudaHostRegisterMapped) ;
    if (cudaSuccess != er)
    {
        printf ("could not register\n") ;
        return 1 ;
    }

    void* d_ptr ;
    er = ::cudaHostGetDevicePointer (&d_ptr, ptr, 0) ;
    if (cudaSuccess != er)
    {
        printf ("could not get device pointer\n") ;
        return 1 ;
    }

    kernel<<<256,256>>> ((int4*)d_ptr) ;

    if (cudaSuccess != ::cudaDeviceSynchronize())
    {
        printf ("error in kernel\n") ;
        return 1 ;
    }

    if (cudaSuccess != ::cudaHostUnregister (ptr))
    {
        printf ("could not unregister\n") ;
        return 1 ;
    }

    ::UnmapViewOfFile (ptr) ;

    ::CloseHandle (hFileMapping) ;
    ::CloseHandle (hFile) ; 

    ::cudaDeviceReset() ;

    printf ("DONE\n");

    return 0 ;
}

【讨论】：

这不只是将所有内容缓存在主内存中并偶尔/最终将更改刷新到磁盘吗？
@einpoklum，在我看来，这对最初的问题来说是一个合理的答案：“基本思想是加载内容（类似于 dma）-> 执行一些操作-> 将内容存储回磁盘（再次在DMA 时尚）。”不过，情况可能确实如此。但是，以编程方式，不需要在 CPU 端实现任何内容。

【解决方案3】：

真正的解决方案即将出现！

抢先体验：https://developer.nvidia.com/gpudirect-storage

GPUDirect® 存储 (GDS) 是 GPUDirect 系列的最新成员。 GDS 为 GPU 内存和存储之间的直接内存访问 (DMA) 传输启用了直接数据路径，从而避免了通过 CPU 的反弹缓冲区。这种直接路径增加了系统带宽并减少了 CPU 上的延迟和利用率负载。

【讨论】：