CUDA 13.3 和 PTX 9.3

CUDA 13.3 不仅仅是一次常规的小版本更新。

明面上的头条自然是“新工具包发布”，但更有看点的故事在于 NVIDIA 正在持续暴露更多的底层机制：tile 级编程、fabric 操作、更丰富的 mbarrier 语义、异步多内存、更好的运行时编译支持，以及一些实实在在的编译器/库修复。

如果你编写内核、构建运行时、调优集合通信，或关注 Blackwell 时代的编程模型，那么这个版本值得一读。

发布概览

CUDA 13.3 包含了更新的编译器、运行时、库和开发者工具组件。

重要版本信息：

• CUDA NVCC：13.3.33
• NVRTC：13.3.33
• CUDA 运行时：13.3.29
• cuBLAS：13.5.1.27
• Nsight Compute：2026.2.0.7
• Nsight Systems：2026.1.3

CUDA 13.x 的次要兼容性仍要求驱动版本 >= 580，但 CUDA 13.3 打包的 Linux 驱动分支为 610.43.02。这一点很重要，因为一些更有趣的平台特性依赖于更新的 r610 栈。

Tile C++ 日益成熟

CUDA Tile C++ 现在已成为该故事中的一等公民。

这是 NVIDIA 在推动更显式的 tile 导向编程模型：更接近内核工程师思考数据移动、矩阵 tile、布局和调度的层次，而无需完全降级到原始的 PTX/SASS。

在 CUDA 13.3 中，Tile C++ 同时支持 nvcc 和 NVRTC。

这一点很重要，因为它使得该模型在预编译和运行时编译工作流中都能发挥作用。

此外还有新的 tile 操作和类型，包括围绕 scaled MMA、pack/unpack 操作、stride 和 gather/scatter 风格的视图、i4 以及一种 4 位浮点类型的支持。

这可能是 CUDA 13.x 中“值得关注”的部分。

C++23 支持正式发布

CUDA 13.3 在 nvcc 和 NVRTC 中增加了对 C++23 的官方支持。

这不是版本中最耀眼的部分，但对于真实的 CUDA 代码库来说意义重大。运行时编译、模板密集型内核、CCCL 使用以及现代 host/device C++ 都将受益于 CUDA 更贴近现代 C++。

NVRTC 头文件处理更完善

NVRTC 现在可以安装并使用经过筛选的 CUDA C++ / CCCL 头文件，而无需像以前那样进行大量手动工具包/头文件配置。

这对于需要动态编译内核的系统来说是一个实际改进：Python 运行时、自动调优器、DSL、模型编译器、推理引擎以及内部内核生成栈。

CUDA Python 跨越重要里程碑

CUDA Python 现在为 cuda.core 和 cuda.compute 提供了稳定的 1.0 API。

cuda.compute 尤其有趣，因为它从 Python 中暴露了基于 CUB 的算法：

• reduce
• scan
• sort
• histogram
• select
• lower / upper bound

这不是“Python 取代 CUDA C++”，而是 CUDA 底层生态系统为工具、原型设计和自动化工作流提供了更好的可编程表面。

PTX 9.3：fabric、mbarrier 和异步多内存

PTX 9.3 才是内核工程师最感兴趣的“糖果”。

新增加的内容主要涉及同步、fabric 级操作、异步多内存以及对内存语义更精细的控制。

新的 fabric 指令：

• fabric.try_get
• fabric.try_put
• fabric.try_red
• fabric.try_pullred
• fabric.wait
• fabric.submit

新的异步多内存操作：

• multimem.st.async
• multimem.red.async

更精细的内存语义控制：

• 异步批量拷贝和 reduce 的 .sem / .scope 限定符
• fence.proxy.to_proxykind::from_proxykind_fabric

更多的 mbarrier 机制：

• .phase_type::*
• reportPredicate
• reportValue
• .layout
• mbarrier.check_layout

其他 PTX 新增内容：

• mma_throughput pragma
• clmad 指令

此外还有语义澄清：red.async 支持 sys scope，而非 cluster scope，且搭配 .release。

重要解读：NVIDIA 正在为多 GPU / fabric 可见行为、异步内存移动和屏障状态搭建更正式的机制。这是运行时作者和编写严肃集合通信的人员应该关注的 PTX 接口。

MPS 获得部分错误隔离

CUDA 13.3 通过静态 SM 分区为 MPS 添加了部分错误隔离能力。

历史上，MPS 下单个客户端的严重故障可能造成比预期更广的影响范围。该特性允许 MPS 将某些 SM 触发的错误隔离到特定分区/客户端。

这并非完美的隔离，NVIDIA 也明确指出了这一点。

但这对于共享 GPU 系统来说仍然是一个重大的运营改进：推理服务、混合在线/离线工作负载、小查询工作负载、机器人/嵌入式栈，以及那些试图在同一 GPU 上保持独立进程运行的系统。

离散 GPU 的 DMA-BUF mmap 支持

CUDA 13.3 为从 CUDA 设备内存导出的 DMA-BUF 文件描述符添加了 mmap() 支持，适用于离散 GPU。

实际含义：在某些环境中，现在有了一条更低延迟的 CPU 映射路径来访问导出的 GPU 内存，而无需完全依赖 GDRCopy 内核驱动。

如果你从事 GPU 内存、CPU 控制代码、I/O 设备或外部内存抽象之间的数据路径工作，这一点值得深入研究。

cuBLAS 在 Blackwell 上获得显著提升

cuBLAS 13.3 包含多个值得注意的改进：

• 绿色上下文支持
• FP64 模拟 GEMM 工作空间上限不超过 8 GB
• Blackwell Ultra 上更好的 FP4 矩阵乘法性能
• Blackwell 和 Blackwell Ultra 上更好的 TF32 矩阵乘法性能
• Hopper 上更好的 TF32 TN 矩阵乘法性能
• Hopper、Blackwell 和 Blackwell Ultra 上基于 TMA 的 SYMV 加速

TF32 性能提升最为突出。NVIDIA 报告称，在 Blackwell / Blackwell Ultra 的问题规模和布局上，几何平均改进达 27%，某些小规模问题甚至提速高达 3.5 倍。

一如既往，请针对你的实际尺寸进行基准测试。

值得注意的正确性修复

有两个正确性修复值得关注。

首先：CUDA 13.3 修复了自 CUDA 12.8 以来存在的一个编译器问题，即编译器插入的线程再聚合在具有多层嵌套发散的核函数中可能失效。这可能导致寄存器值陈旧或损坏，从而引发错误执行。

这类 Bug 是内核工程师应该立即关注的。

其次：CUDA Math 修复了一个涉及 __mul24() 和编译时常量输入的静默数据损坏问题。该问题自 CUDA 11.1 引入，在 CUDA 13.3 中得到解决。

已知注意事项

几点需要注意：

• 不再提供传统的 Nsight Eclipse Edition 插件。
• CUDA Tile C++ 存在一个已知问题：当 GLIBC 强化级别 2+ 启用时，tile 函数中的 printf() 可能产生错误的诊断信息。解决方法：添加 -U_FORTIFY_SOURCE。
• Compute Fabric Transport 存在一些已知问题。单播逻辑端点需要 Fabric Manager 610+，fabric.try_pullred 目前并非最优，某些逻辑端点内存绑定在同进程多 GPU 指针访问时可能表现不正确。

总结

CUDA 13.3 之所以有趣，是因为它延续了 CUDA 13.x 的趋势：对底层机制更显式的控制、更丰富的 tile/fabric 级编程接口、更好的运行时编译体验，以及 Blackwell 时代有意义的库调优。

对于大多数应用程序开发者来说，这是一次不错的工具包更新。

对于内核工程师、编译器人员、运行时作者以及任何在多设备上构建 GPU 系统的人来说，PTX 9.3 是需要仔细阅读的部分。

本文由 GPT 5.5 Medium Fast 撰写