GPUDirect Storage 实战:让数据从 NVMe 直达 GPU 显存
一句话:传统读盘要经过"磁盘 → CPU 内存 bounce buffer → GPU 显存"两次拷贝;GPUDirect Storage(GDS)让存储设备通过 DMA 直接把数据搬进 GPU 显存,CPU 只负责发命令不负责搬数。网络侧的直通叫 GPUDirect RDMA,存储侧的直通就是 GDS——一个管东西怎么进来,一个管东西怎么出去。
集群里 GPU 算得越来越快,I/O 反而成了越来越常见的瓶颈:训练启动时要加载 TB 级数据集,训练中要周期性写 checkpoint,推理服务要不断从磁盘换模型权重。这篇文章讲透 GDS 的数据路径、生效条件、压测方法和接入方式,以及最容易踩的"以为开了直通、其实在 compat 模式跑"的坑。
一、先看传统路径:两次拷贝,CPU 全程当搬运工
没有 GDS 时,一个读请求的数据流是这样的:
NVMe / 网络存储
│ DMA(写入内核页缓存 / 用户态缓冲区)
▼
CPU 系统内存 ← 第一次落地:存储到内存
│ cudaMemcpy(CPU 发起、占用内存带宽与 CPU 周期)
▼
GPU 显存 ← 第二次落地:内存到显存
▼
CUDA kernel 开始计算这条路径有三个代价:
- 数据两次落地,延迟天然多一跳;
- CPU 内存带宽被占。多路 NVMe 顺序读可以打到几十 GB/s,这些流量全部要过 CPU 的内存控制器,和 dataloader 的预处理进程抢带宽;
- CPU 占用率虚高,大量周期花在内存拷贝上,而不是数据解码和增广。
GDS 的改法是把中间那层 bounce buffer 拿掉:
NVMe / 支持 RDMA 的存储网卡
│ 点对点 DMA(P2P DMA,地址由 nvidia-fs / P2PDMA 翻译)
▼
GPU 显存 ← 一次落地
▼
CUDA kernel 开始计算DMA 引擎本来就会做地址翻译,传统操作系统之所以做不到,是因为它只会把数据搬进内核管理的物理页,不认识 GPU 的虚拟地址。GDS 补上的就是这一环:文件系统驱动在构造 DMA 散列聚集表(scatter-gather list)时,通过回调拿到 GPU 虚拟地址对应的物理地址,DMA 就能直达显存。
二、软件栈三件套:libcufile、nvidia-fs.ko 与新的 P2PDMA 模式
GDS 不是一个单一驱动,而是一组组件:
| 组件 | 位置 | 职责 |
|---|---|---|
| libcufile.so | 用户态 | 应用面对的 cuFile API(cuFileRead/cuFileWrite 等),负责路径选择、缓冲区管理与批量提交 |
| nvidia-fs.ko | 内核态(CUDA 12.8 之前的主路径) | 接收用户态 IOCTL,向存储驱动提供 DMA 回调,把 GPU 虚拟地址翻译成物理地址 |
| 兼容模式(compat) | 用户态 | nvidia-fs 不可用时,退回 POSIX pread/pwrite 到系统内存再拷进显存,功能正确但没有直通收益 |
CUDA 12.8 之后多了一条更"干净"的路径:x86_64 平台可以直接使用 Linux 内核上游的 PCI P2PDMA 基础设施,不再依赖 nvidia-fs.ko,也不需要给 nvme.ko 打定制补丁。启用条件是内核版本足够新(Ubuntu 系通常要求 6.2 以上)、OpenRM 驱动 570.x 以上并打开对应注册键。要注意两个限制:上游 P2PDMA 目前不支持 RAID0 和 NVMe 多路径,这两种形态仍要走 nvidia-fs 路径。可以用下面命令判断当前内核是否具备 P2PDMA 能力:
cat /proc/kallsyms | grep -i p2pdma_pgmap_ops
# 有输出即具备上游 P2PDMA 支持三、直通成立的条件:不是装了包就一定走 GDS
GDS 的"开了"和"生效"是两回事。一次 I/O 能不能走直通,同时取决于存储设备、PCIe 拓扑、文件系统和对齐方式。
3.1 文件系统与存储形态支持矩阵
| 存储形态 | GDS 支持情况 | 备注 |
|---|---|---|
| 本地 NVMe + ext4 / xfs | 支持,最成熟的场景 | O_DIRECT 路径,4KB 对齐 |
| NVMe-oF(RDMA) | 支持 | 远端 NVMe 经 RDMA 网卡直达显存 |
| Lustre | 支持 | 小 I/O 下 POSIX 路径延迟可能更低,可用 posix_gds_min_kb 设阈值 |
| IBM Spectrum Scale(GPFS) | dmabuf / compat 路径 | 按发行版与客户端版本能力不同 |
| BeeGFS | 支持,但需要专门的客户端配置 | 未配置时 gdscheck 会显示 compat |
| WekaFS、VAST/DDN 等 NFS 系方案 | 支持 | 依赖厂商插件与动态路由 |
| 普通 NFS / 不支持的文件系统 | 自动回退 compat | 功能可用,无直通收益 |
3.2 拓扑与硬件条件
- GPU 与 NVMe(或存储网卡)尽量在同一个 PCIe switch / 同一 root port 下。跨 CPU root complex 的 P2P 流量延迟更高,此时 cuFile 的动态路由会改走"最近的 GPU 当中转 + NVLink 二次拷贝"的方案,仍然优于纯 CPU 路径,但不是最优;
- PCIe ACS(Access Control Services) 如果在中间交换机上被强制开启,会阻断 P2P 事务,
gdscheck会明确报出是哪一级 switch; - BAR1 显存映射空间要足够。老卡或未开启 Resizable BAR 的平台上,可映射的显存窗口受限,超出部分的 I/O 会走 bounce buffer;
- 远端存储要求网卡具备 RDMA 能力(IB 或 RoCE),否则网络路径无法 P2P;
- 4KB 对齐是硬要求:文件偏移和 I/O 大小都要按页对齐,非对齐写会自动退回 compat。
3.3 一个容易被忽略的边界:托管内存
cudaMallocManaged 分配的可迁移内存、系统 malloc/栈内存,在 GDS P2P 模式下不被支持,cuFile 会改用内部 bounce buffer 的兼容路径完成。直通的收益只对 cudaMalloc / cuMemAlloc / cuMemCreate 分配的设备显存成立。
四、compat 模式:功能不报错,但性能是假的
兼容模式是 GDS 最容易误导人的设计:应用调用 cuFile API 全部成功返回,但数据实际走的是 POSIX + CPU 拷贝。默认配置 allow_compat_mode: true,库在检测到无法直通时静默回退,不会报错中断。
官方文档列出的典型回退触发条件:
- 硬件配置不在支持列表内;
- GPU 计算能力达标但没有足够 BAR 空间暴露;
nvidia-fs.ko没有加载或加载失败;- Linux 发行版不在支持范围;
- 文件系统不被 GDS 支持;
- 网络链路没有启用 RDMA,或配置未达到直通最优条件;
- 非 4KB 对齐的写 I/O。
所以验收 GDS 的核心动作不是"应用跑通了",而是证明流量真的走了 nvfs 直通路径。两种验证手段:
# 1. 打开 cuFile 三级统计,nvfs 计数增长才是直通,posix 增长说明在 compat
gds_stats -p <pid> -l 3
# 2. KvikIO/Python 场景同样看 gds_stats,区分 nvfs 与 posix 两个来源如果统计全部落在 posix 一栏,先按上面七条逐一排查,而不是怀疑应用代码。
五、安装与上线前体检
RHEL/Rocky 系通过仓库安装三个包(Ubuntu 系包名一致,用 apt):
# nvidia-gds 是元包,会带上内核模块与用户态库;gds-tools 提供 gdsio/gdscheck/gds_stats
dnf install -y nvidia-gds nvidia-fs-dkms gds-tools
# 加载内核模块(P2PDMA 新模式下可不依赖它,但传统路径必需)
modprobe nvidia-fs
lsmod | grep nvidia_fs装完先用 gdscheck 做两级体检,不要急着上业务:
# 平台级:GPU、驱动、nvidia-fs、PCIe ACS、BAR 空间、RDMA 设备是否就绪
gdscheck -p
# 挂载级:每个文件系统挂载点被识别为 supported 还是 compat
gdscheck -mgdscheck -m 的输出要逐行看:同样是并行文件系统客户端,挂载参数不对时会从 supported 掉成 compat,这一步能在业务上线前把问题拦住。
六、gdsio 压测:用同一套参数对照两条路径
gds-tools 自带的 gdsio 是 GDS 版的 fio,专门测存储到 GPU 显存的传输。关键参数:
| 参数 | 含义 |
|---|---|
-f / -D | 目标文件 / 目标目录(目录模式配合 -w 批量写多文件) |
-d | 数据落点的 GPU 编号 |
-s | 数据集大小 |
-i | 单次 I/O 大小(判断大块/小块收益的关键变量) |
-x 0 | GDS 直通路径(GPUD) |
-x 2 | 传统 POSIX 经 CPU 中转(r1.12;老版本该值为 1,以 gdsio -h 为准) |
-I 1 | 写;读用默认/0 |
-w | I/O 线程数 |
-T | 持续时间(秒) |
一组最小对照实验:
# 直通路径:1GiB 文件、1MiB 的 I/O、8 线程、读到 GPU 0
gdsio -f /mnt/nvme/gds.test -d 0 -s 1G -i 1M -w 8 -x 0
# 同参数走传统 CPU 中转路径
gdsio -f /mnt/nvme/gds.test -d 0 -s 1G -i 1M -w 8 -x 2读输出时看三个字段:Throughput(聚合吞吐)、Avg_Latency(平均延迟)和 ops。官方文档给过一个双 NVMe RAID0 的实测样例:单盘顺序读约 3.4 GB/s 的两块盘,gdsio 直通读拿到约 6.5 GiB/s 的持续吞吐、平均延迟约 1.2 ms,基本把硬件带宽吃满。
对照实验必须同时量化 CPU 代价。压测期间另开终端跑:
iostat -cxzk 1重点看两点:一是直通模式下 NVMe 设备 rkB/s 打满而 %system 明显更低;二是 compat 模式下 user/system 占用显著抬高。只比吞吐不比 CPU,会低估 GDS 在大规模 dataloader 并发时的真实价值——省下的 CPU 周期本来是要留给数据解码的。
小块 I/O 不要想当然
Lustre 这类网络文件系统上,I/O 太小的时候 POSIX 路径的延迟反而可能更低(P2P 建连与地址翻译的固定成本摊不薄)。cufile.json 里的 lustre:posix_gds_min_kb 就是用来设这个分流阈值的:小于阈值走 POSIX,大于阈值走 GDS。压测时要扫一遍 -i 从 4K 到 1M 的曲线,用拐点而不是信仰来配这个值。
七、应用怎么接:从 cuFile 到 PyTorch
7.1 cuFile 原生调用序列
C/C++ 应用直接用 cuFile API,生命周期非常固定:
cuFileDriverOpen() # 进程级初始化
cuFileBufRegister(gpuPtr, size) # 注册 GPU 缓冲区(可复用)
cuFileHandleRegister(fd) # 打开并注册文件句柄
cuFileRead / cuFileWrite # 直通读写,可挂 CUDA stream 做异步
cuFileHandleDeregister()
cuFileBufDeregister()
cuFileDriverClose()编译链接 -lcufile -lcuda -lcudart,头文件在 /usr/local/cuda/include/cufile.h。Stream 子集 API 支持把 I/O 排进 CUDA 流,实现"kernel 产出数据 → I/O 落盘"或"I/O 完成 → kernel 启动"的严格序,多个流还能并发吃满多个 DMA 引擎。
7.2 PyTorch 2.7:torch.cuda.gds 原型 API
PyTorch 从 2.7 开始在 torch.cuda.gds 提供 cuFile 的薄封装(要求 CUDA 12.6+),目前标注为 prototype,最直接的用途是加速 torch.save / torch.load 的 checkpoint 路径:
GdsFile是文件对象封装,打开时自动带上O_DIRECT;save_storage/load_storage分别绑定同步的 cuFileWrite / cuFileRead;gds_register_buffer包装 cuFileBufRegister,反复读写同一块显存时只注册一次;- GDS 要求 4KB 对齐,PyTorch 侧通过
torch.utils.serialization.config.save.storage_alignment配置。
大模型 checkpoint 单文件往往几十上百 GB,正是大块顺序 I/O 最吃 GDS 红利的场景。但因为 API 仍是原型,生产上建议先在影子节点验证正确性与回退行为,再随 PyTorch 版本升级推进。
7.3 Python 数据栈:KvikIO 与 DALI
- KvikIO 是 CUDA-X 数据科学栈(cuDF、cuCIM 等)使用的 Python I/O 层,底层调 cuFile,Parquet/ORC 列式加载、病理图像读取都走它。注意 KvikIO 和 cuFile 是两层决策:KvikIO 决定要不要调 cuFile,cuFile 决定最终走 nvfs 还是 posix,所以验证仍以 gds_stats 为准;
- DALI 的文件读取管线提供 GDS 开关,适合图像类训练把"磁盘解码前的原始读"也移出 CPU;
- 自行开发的 dataloader 不必直接啃 cuFile C API,优先评估 KvikIO,把路径判定和 compat 回退交给成熟库。
八、动态路由:GPU 和存储网卡不在一个 PCIe 域怎么办
网络文件系统场景下,GPU 与存储网卡不共享 root port 时,硬做跨 root complex 的 P2P 反而更慢。cuFile 的动态路由会为每个挂载点预算"离存储网卡 PCI 距离最近的 GPU":
- 目标 GPU 与网卡在同一 PCIe switch 下 → 直接 P2P,不触发路由;
- 不在同一 switch → 选一张最近的 GPU 作 bounce,先 P2P 到该 GPU,再通过 NVLink 做设备间拷贝到目标 GPU;
- 全程不经过 CPU 系统内存,CPU 仍然只发命令。
这解释了一个现场现象:换了 GPU 槽位后 gdsio 吞吐变化明显。先看 gdscheck -p 的拓扑距离,再决定网卡与 GPU 的插槽配对,而不是先怀疑文件系统。
九、五个最常见的坑
- 装了 nvidia-gds 就当直通生效。实际上
allow_compat_mode默认静默回退,必须用 gds_stats 看到 nvfs 计数才算数; - 忽略 4KB 对齐。偏移或 I/O 长度不对齐,写请求全部悄悄走 POSIX,应用层毫无报错;
- RAID0 / NVMe 多路径上强开 P2PDMA 新模式。上游内核暂不支持,会直接失败或回退,这类形态仍需 nvidia-fs 路径;
- 只看吞吐不看 CPU 占用。GDS 的一半收益是把 CPU 从拷贝里解放出来给 dataloader,不做 iostat 对照等于只评估了一半;
- 小块 I/O 一刀切走 GDS。Lustre 等网络文件系统上小 I/O 的固定开销摊不薄,要用
posix_gds_min_kb按实测拐点分流。
十、上线验收清单
lsmod | grep nvidia_fs模块在位(P2PDMA 模式除外),gdscheck -p无 ACS 阻断、BAR 与 RDMA 设备就绪;gdscheck -m目标挂载点逐行显示 supported,而不是 compat;- gdsio 用
-x 0与传统路径同参数对照,吞吐、延迟、iostat 的 CPU 占用三组数据留档; - 扫一遍 4K → 1MiB 的 I/O 大小曲线,确定小 I/O 分流阈值;
- 业务进程跑起来后用
gds_stats -p <pid> -l 3确认 nvfs 计数增长、posix 不增长; - PyTorch checkpoint 场景验证 GDS 写出的文件可被普通
torch.load正确读回,回退路径可用; - 故障注入一次:卸载 nvidia-fs 或换一个不支持的挂载点,确认应用能安全回退 compat 而不是崩溃。
收口三句
- GDS 的本质是省掉 CPU 内存这一跳:DMA 直达显存,收益同时体现在带宽、延迟和 CPU 占用三个维度;
- "装了"不等于"通了":支持矩阵、拓扑、对齐任一不满足都会静默回退 compat,gds_stats 的 nvfs/posix 计数是唯一可信的判官;
- 大块顺序 I/O 收益最大,小块随机 I/O 要测拐点:checkpoint 和数据集预加载是主战场,分流阈值交给压测数据而不是经验。