OpenDLSS:英伟达 DLSS 5 神经渲染网络的 Vulkan 重新实现

Hacker News Top 工具

摘要

OpenDLSS-NR 是 NVIDIA DLSS 5 生成式神经渲染网络(71 层 Swin/ViT U-net)的开源、逐比特一致的 Vulkan 重新实现,另有一个独立的 WebGPU 移植版本,可在没有张量核心和 FP8 的情况下于浏览器中运行。

暂无内容
查看原文
查看缓存全文

缓存时间: 2026/10/01 11:09

maanHimself/OpenDLSS-NR 来源:https://github.com/maanHimself/OpenDLSS-NR

OpenDLSS-NR

NVIDIA 的 DLSS 5 神经渲染网络的 Vulkan 重新实现,与原始实现逐位一致。与 DLSS-NR build 310.8.0 使用完全相同的 71 层 Swin / ViT 网络,在 Tensor Core 上以 FP8 运行。中间结果同样一致,不只是最终输出:全部 75 个层级边界,逐字节比对。

ports/browser-webgpu/ 是第二个独立实现:在浏览器中运行相同字节的网络,没有 Tensor Core,也不使用 FP8。

你自行提供权重,按下述布局组织成一个模型目录。

网络

这是一个 U-net,由移位窗口 Transformer 块构成,底部是一个全局 ViT:跨六个池化层级共 71 个块,FP8(E4M3)激活值配合 FP16 累加,权重共 141 MiB。这是一个生成式神经渲染网络(NVIDIA 的说法):它会重新渲染引擎已经画好的帧,从注入的噪声中生成细节,并在风格设置下调整色调、结构和皮肤。输入和输出分辨率相同;它不是超分辨率网络。

同一帧关闭(左)和开启(右)神经渲染的对比

WebGPU 移植版在 2048x1152 下的对比,左侧关闭 NR,右侧开启。场景:Cowboy Gramps(https://www.blendkit.com/asset-gallery-detail/96dce188-9c9c-4699-a45a-48663fbbbcb7/),作者 Muhammed Ismayil,CC0 协议。

它接收一帧已渲染的图像(该帧的低动态范围代理、三路高斯噪声、上一帧输出的重投影结果以及五个条件标量),并输出每个像素四个 f32 通道:一个 RGB 残差和一个时域混合 logit。docs/network.md 是完整的网络图。NVIDIA 在其报告《DLSS 5: Generative Neural Rendering》(https://research.nvidia.com/labs/adlr/DLSS5/files/DLSS5_Report.pdf)(项目主页:https://research.nvidia.com/labs/adlr/DLSS5/)中对该模型有描述。

构建与运行

powershell -File scripts\fetch_tools.ps1 [-Npm] # 首次:tools\(glslang、Vulkan-Headers、volk、CMake、Ninja)
powershell -File scripts\build.ps1 # 着色器、PTX、build\dlss5vk.exe
powershell -File scripts\fetch_filament.ps1 # 首次,用于演示:third_party\filament(以及补丁)
powershell -File scripts\build_filament.ps1 # 首次,用于演示:third_party\filament-install
powershell -File scripts\build_demo.ps1 # build\demo\dlss5-demo.exe
build\dlss5vk.exe bench --model <模型目录> --width 768 --height 768
build\dlss5vk.exe profile --model <模型目录> --width 768 --height 768 # 逐 dispatch 计时
build\dlss5vk.exe parity --model <模型目录> --fixture <基准数据目录> # 与基准数据逐位一致性比对
build\dlss5vk.exe verify --model <模型目录> --fixture <基准数据目录> # block-0 的逐 kernel 二分定位
python scripts\ptx\test_fast_divmod.py # PTX 除法器,遍历所有 n < 2^24(numpy)

演示程序可以直接双击运行。它会列出 build\scenes 下的每个场景,显示在 Demo scene 下拉菜单中并从第一个场景开始,或者加载命令行中给出的 glTF 文件。模型目录的优先级为 --model <路径>,其次是环境变量 DLSS5VK_MODEL,最后是本 README 旁的 models\nr。渲染器、按键、场景和 view.json 请参见 demo/README.md。

性能

RTX 4070 SUPER,整个网络每帧耗时,40 帧取最小值。所有分辨率下均为 241 个 dispatch。

分辨率耗时
768x7682.8 ms
1920x10807.8 ms
2560x144012.6 ms
3840x216029.3 ms

GPU 在持续负载下会交替使用两种时钟状态,因此中位数会高出百分之几。请以最小值进行比较。

内容构成

组件文件说明
主机端src/(C++20)Vulkan 上下文、模型加载与权重重排、kernel 封装、网络图、算术运算的 CPU 参考实现、dlss5vk 工具
GLSL kernelshaders/参考路径:cooperative-matrix FP8 GEMM、融合的 32 通道 block、融合的 QKV + 窗口注意力、专家 MLP、全局注意力、逐元素运算。本身完整且精确。
PTX kernelscripts/ptx/生成 PTX 的 Python 生成器,对应快速路径:mma.sync E4M3 配合 f16 累加、cp.async 环形缓冲、通过设备计数器实现无 barrier 串联、split-K GEMM、流式全局注意力。由构建过程生成到 build/ptx。
演示demo/、third_party/filament.patch将网络置于 Filament(Apache-2.0)框架中:为 Filament 打补丁以支持逐物体运动矢量和 Vulkan 互操作钩子,通过 gltfio 加载 glTF 场景,ImGui 控件。
WebGPU 移植ports/browser-webgpu/在浏览器中运行相同的网络,与相同的采集数据逐位一致,没有 Tensor Core、没有 FP8、块之间没有融合、没有串联:精确性来自规范本身,而非硬件。512x512 下需 72 ms,而这里只需 2.7 ms。

未实现:DLSS-SR,那是另一个不同的网络。时域路径已实现,但仅在演示中:网络的历史输入通道和逐像素混合 logit 驱动一个重投影反馈循环(docs/frame.md)。dlss5vk 工具运行的是无历史记录的单帧,参考采集数据正是用这种方式生成的。

环境要求

  • Windows、NVIDIA Ada(或更新)GPU,以及暴露 VK_KHR_cooperative_matrix、VK_NV_cooperative_matrix2、VK_EXT_shader_float8 和 VK_NV_cuda_kernel_launch 的驱动。
  • Visual Studio 2022 或更高版本,含 C++ x64 工具集(任意版本或仅安装 Build Tools;通过 vswhere 定位,或将 VCVARS64 设为你的 vcvars64.bat)、git、用于 PTX 生成器的 Python 3,以及用于场景转换器的 Node.js + npm 和 Pillow。
  • tools/ 下的便携工具链(已加入 .gitignore):scripts\fetch_tools.ps1 会下载 glslang 16.6.0、Vulkan-Headers v1.4.363、volk(固定标签)、CMake 3.31 和 Ninja 1.13。无需安装 Vulkan SDK。-Npm 还会把场景转换器的依赖模块安装到 tools\gltf。
  • 演示所需:Filament v1.77.0,由 scripts\fetch_filament.ps1 克隆并打补丁,再由 scripts\build_filament.ps1 构建一次(两者均已在 .gitignore 中;大约 15 分钟,构建目录约 6 GB,位于 %LOCALAPPDATA%\dlss5-vulkan 或 DLSS5_FILAMENT_BUILD_DIR;DLSS5_BUILD_JOBS 可限制并行编译数,默认 8,因为 MSVC 在 Filament 上每个任务最多占用约 1 GB 内存)。

模型目录

nr::Model 读取 manifest.json:一个 stages 数组(每项含 id、相对于目录的 file、packedByteLength、sha256)和一个 tensors 数组(每项含 name、block、layer、parameter、stage、stageOffset、byteLength)。Stage 文件以打包字节形式存放 E4M3 权重;主机会将其重新布局为 kernel 消耗的矩阵形式(src/nr_model.cpp)。本仓库中没有任何代码能生成这样的目录。该图就是 310.8.0 的 71 块网络,别无其他:块数量不同的模型会在加载时被拒绝。

基准数据(Fixtures)

parity 与原始实现的录制采集数据进行比对,这些数据不属于本仓库。一个 fixture 是包含 manifest.json 的目录,字段如下:

键说明
sourceDimensions、fullDimensions有效尺寸与补零后的尺寸
proxy 或 inputFeatures输入:RGBA f32 图像(含 conditioning、seed、autoMask),或直接是 f32 特征
checks该 fixture 校验的内容,取值为 "boundaries"、"head"、"output" 中的任意一个;必填且不能为空
blocks、transitions"boundaries":E4M3 参考数据(block / id、width、height、channels、file)
omittedBoundaries"boundaries":每个该 fixture 缺少参考数据的可比边界对应的 {name: reason}
referenceHead"head":f32 RGBA head
nativeOutput"output":合成后的图像,dtype 为 "f32"(RGBA 半精度,需要 proxy)或 "u8"(8 位采集数据)

所有内容都会在 GPU 运行前完成校验,任何一项不通过的 fixture 都会被拒绝:声明了某个校验项却缺少对应参考数据、参考数据缺失、长度不足或未指向图中的任何对象、参考数据没有被任何已声明的校验项使用,或者某个可比边界(block 0-69 以及五个编码器 transition)既没有参考数据也没有给出原因。

判定结果为 bit-exact(通过)、仅零的符号不同(失败)、在一个码值以内(仅限 8 位采集数据,单独报告)或不匹配;详见 docs/numerics.md。head 和 output 按生产调度进行比较,会以 --repeat 次(默认 3)重新提交,且必须与带 barrier 的相同网络结果一致;边界数据来自一次插桩运行,其 head 必须与生产运行的结果一致。verify 还需要 inputFeatures 和一个 block-0 参考数据。

调优开关

所有开关默认使用快速且精确的路径。每个开关都保证输出逐字节相同,且在每个开关下的 parity 都是门禁的一部分。任何将 kernel 退回 GLSL 路径的开关也会同时关闭计数器串联,因为只有 PTX kernel 参与其中。

  • DLSS5VK_UNFUSED=1 使用 GLSL 参考路径,最大支持 2560x1440:会实例化每一个中间结果。
  • DLSS5VK_PTX_DIR 指定 PTX 目录,默认 build/ptx。
  • DLSS5VK_CHAIN=0 在每次启动之间插入 barrier,替代计数器串联。
  • DLSS5VK_PTX_GEMM、GEMMT、GEMMV、BLOCK32、FFN、QKV 和 ATTN 设为 0 会将对应的 kernel 家族退回 GLSL 写法。
  • DLSS5VK_ATTN_STREAM 强制关闭(0)或开启(1)流式全局注意力。
  • DLSS5VK_SPLITK=0、DLSS5VK_VIT_CHAIN=0 和 DLSS5VK_NO_PTX_MLP=1 分别禁用 split-K、ViT 串联和 PTX MLP。
  • DLSS5VK_NO_FUSE_PRE、POOL、UPRES 和 POST 设为 1 会各自取消一处融合。
  • DLSS5VK_CHAIN_MASK 是位掩码:1 专家级 stage,2 c32 block,4 split GEMM。默认 3。
  • DLSS5VK_DEFER_MAX 是投影 GEMM 被融合的最大 stage 宽度。默认 128。
  • DLSS5VK_VALIDATION=1 在 Khronos 验证层下运行(若未安装则拒绝运行:需要 Vulkan SDK,或用 VK_LAYER_PATH 指向 Vulkan-ValidationLayers 的构建目录);验证层报告的任何错误都会导致运行失败。DLSS5VK_DEBUG=1 仅打印驱动自身的消息(其中包括 PTX 编译器的消息)。DLSS5VK_LIST_EXTENSIONS=1 列出扩展。

文档

docs/README.md 是索引:network.md 是网络图,numerics.md 是精确性契约,weights.md 是权重布局,execution.md 是调度,frame.md 是演示中的帧流程。

与 NVIDIA 无关联

本项目不隶属于、未获 NVIDIA 认可、也不受 NVIDIA 支持。其中不含任何 NVIDIA 软件、权重、头文件或获取这些内容的说明。本仓库及其许可证不授予、也不暗示授予任何 NVIDIA 知识产权下的权利;你需自行负责与你所使用的模型数据相关的许可证。

许可证

本仓库中的一切均采用 MIT 许可证(LICENSE)。第三方组件列于 NOTICE。

相似文章

Nvidia DLSS 5 详解

The Verge

本文解释了Nvidia的DLSS 5,这是一项用于电子游戏的生成式AI技术,能增强光照和细节,回应了过去的批评,并在九月发布前强调了开发者的控制权。