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

Posted 2023-03-16

【中文标题】可以直接从gpu访问硬盘吗？

【英文标题】：Is it possible to access hard disk directly from 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 链接的一半到系统。它可能以较低的延迟这样做，但在我看来并不像游戏规则改变者。

【参考方案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 上的延迟和利用率负载。