剛剛，英偉達CUDA迎來史上最大更新！

機器之心報道

機器之心編輯部

幾個小時前，NVIDIA CUDA Toolkit 13.1 正式發(fā)布，英偉達官方表示：「這是 20 年來最大的一次更新。」

這個自 2006 年 CUDA 平臺誕生以來規(guī)模最大、最全面的更新包括：

• NVIDIA CUDA Tile 的發(fā)布，這是英偉達基于 tile 的編程模型，可用于抽象化專用硬件，包括張量核心。
• Runtime API exposure of green contexts（是指把所謂的 Green Context「指輕量級的、可并發(fā)調(diào)度的上下文或執(zhí)行環(huán)境」暴露給外部調(diào)用者使用。）
• NVIDIA cuBLAS 中的雙精度和單精度仿真。
• 一本完全重寫的 CUDA 編程指南 ，專為 CUDA 新手和高級程序員設計。

下面我們就來具體看看。

CUDA Tile

CUDA Tile 是 NVIDIA CUDA Toolkit 13.1 最核心的更新。它是一種基于 tile 的編程模型，能夠以更高的層次編寫算法，并抽象化專用硬件（例如張量核心）的細節(jié)。

解讀 CUDA Tile 的核心概念

英偉達博客解釋說：CUDA Tile 可讓開發(fā)者在高于 SIMT（單指令多線程）的層級編寫 GPU 核函數(shù)。

在目前的 SIMT 編程中，開發(fā)者通常通過劃分數(shù)據(jù)并定義每個線程的執(zhí)行路徑來指定核函數(shù)。

而借助 CUDA Tile，開發(fā)者可以提升代碼的抽象層級，直接指定被稱為「Tile」的數(shù)據(jù)塊。只需指定要在這些 Tile 上執(zhí)行的數(shù)學運算，編譯器和運行時環(huán)境會自動決定將工作負載分發(fā)到各個線程的最佳方式。

這種 Tile 模型屏蔽了調(diào)用 Tensor Core 等專用硬件的底層細節(jié)，并且 Tile 代碼將能夠兼容未來的 GPU 架構(gòu)。

CUDA 13.1 包含兩個用于 Tile 編程的組件：

• CUDA Tile IR：一種用于 NVIDIA GPU 編程的全新虛擬指令集架構(gòu)（ISA）。
• cuTile Python：一種新的領(lǐng)域特定語言（DSL），用于在 Python 中編寫基于數(shù)組和 Tile 的核函數(shù)。

編譯的 Tile 路徑可以融入完整的軟件棧，與 SIMT 路徑對應。

這是該軟件的首個版本，其包含以下注意事項：

• CUDA Tile 僅支持 NVIDIA Blackwell（計算能力 10.x 和 12.x）系列產(chǎn)品。未來的 CUDA 版本將擴展對更多架構(gòu)的支持。
• 目前的開發(fā)重點聚焦于 AI 算法的 Tile 編程。英偉達表示在未來的 CUDA 版本中將持續(xù)增加更多特性、功能并提升性能。
• 英偉達計劃在即將發(fā)布的 CUDA 版本中引入 C++ 實現(xiàn)。

為什么要為 GPU 引入 Tile 編程？

CUDA 向開發(fā)者提供了單指令多線程（SIMT）硬件和編程模型。這種模式要求（同時也允許）開發(fā)者以最大的靈活性和針對性，對代碼的執(zhí)行方式進行細粒度控制。然而，編寫高性能代碼往往需要付出巨大的心力，尤其是在需要適配多種 GPU 架構(gòu)的情況下。

盡管已有許多庫（如 NVIDIA CUDA-X 和 NVIDIA CUTLASS）旨在幫助開發(fā)者挖掘性能，但CUDA Tile 引入了一種比 SIMT 層級更高的新型 GPU 編程方式。

隨著計算工作負載的演進，特別是在 AI 領(lǐng)域，張量已成為一種基礎(chǔ)數(shù)據(jù)類型。NVIDIA 開發(fā)了專門用于處理張量的硬件，例如 NVIDIA Tensor Core（TC）和 NVIDIA Tensor Memory Accelerator（TMA），它們現(xiàn)已成為每個新 GPU 架構(gòu)中不可或缺的組成部分。

硬件越復雜，就越需要軟件來幫助駕馭這些能力。CUDA Tile 對 Tensor Core 及其編程模型進行了抽象，使得使用 CUDA Tile 編寫的代碼能夠兼容當前及未來的 Tensor Core 架構(gòu)。

基于 Tile 的編程方式允許開發(fā)者通過指定數(shù)據(jù)塊（即 Tile），然后定義在這些 Tile 上執(zhí)行的計算來編寫算法。開發(fā)者無需在逐元素的層面上設定算法的執(zhí)行細節(jié)：編譯器和運行時將處理這些工作。

下圖展示了隨 CUDA Tile 推出的 Tile 模型與 CUDA SIMT 模型之間的概念差異。

Tile 模型（左）將數(shù)據(jù)劃分為多個塊，編譯器將其映射到線程。單指令多線程（SIMT）模型（右）將數(shù)據(jù)同時映射到塊和線程

這種編程范式在 Python 等語言中很常見，在這些語言中，像 NumPy 這樣的庫可以讓開發(fā)者指定矩陣等數(shù)據(jù)類型，然后用簡單的代碼指定并執(zhí)行批量操作。

CUDA 軟件更新

以下是本次 CUDA 版本更新中包含的其他重要軟件改進：

運行時對 Green Context（綠色上下文）的支持

CUDA 中的 Green Context 是一種輕量級的上下文形式，可作為傳統(tǒng) CUDA 上下文的替代方案，為開發(fā)者提供更細粒度的 GPU 空間劃分與資源分配能力。

自 CUDA 12.4 起，它們已在驅(qū)動 API 中提供；而從本版本開始，Green Context 也正式在運行時 API 中開放使用。

Green Context 使用戶能夠定義和管理 GPU 資源的獨立分區(qū)，主要是 Streaming Multiprocessors（SM）。你可以將特定數(shù)量的 SM 分配給某個特定的 Green Context ，然后在該 context 所擁有的資源范圍內(nèi)啟動 CUDA kernel 并管理只在此 context 內(nèi)運行的 stream。

一個典型的應用場景是：你的程序中有部分代碼對延遲極為敏感，并且需要優(yōu)先于其他所有 GPU 工作執(zhí)行。通過為這段代碼單獨創(chuàng)建一個 Green Context 并分配 SM 資源，而將剩余的 SM 分配給另一個 Green Context 處理其他任務，你就能確保始終有可用的 SM 供高優(yōu)先級計算使用。

CUDA 13.1 還引入了更加可定制的 split () API。開發(fā)者可以通過這一接口構(gòu)建此前需要多次 API 調(diào)用才能完成的 SM 分區(qū)，并且可以配置工作隊列，從而減少不同 Green Context 之間提交任務時產(chǎn)生的偽依賴（false dependencies）。

有關(guān)這些功能及 Green Context 的更多信息，請參見 CUDA Programming Guide。

• CUDA 編程指南地址：https://docs.nvidia.com/cuda/cuda-programming-guide/04-special-topics/green-contexts.html

CUDA 多進程服務（MPS）更新

CUDA 13.1 為多進程服務帶來了多項新特性和功能。有關(guān)這些新功能的完整信息，請參閱 MPS 文檔。以下是部分亮點內(nèi)容：

內(nèi)存局部性優(yōu)化分區(qū)

內(nèi)存局部性優(yōu)化分區(qū)（Memory locality optimization partition，MLOPart）是 NVIDIA Blackwell 系列（計算能力 10.0 和 10.3，為架構(gòu)版本號）及更新 GPU 上提供的一項特性。

該功能允許用戶創(chuàng)建專門優(yōu)化內(nèi)存局部性的 CUDA 設備。MLOPart 設備基于同一塊物理 GPU 派生而來，但呈現(xiàn)為多個獨立設備，每個設備擁有更少的計算資源和更小的可用內(nèi)存。

在計算能力 10.0 和 10.3 的 GPU 上，每塊 GPU 都包含兩個分區(qū)。

當在 GPU 上啟用 MLOPart 時，每個分區(qū)都會作為一個獨立的 CUDA 設備出現(xiàn)，并具有其對應的計算與內(nèi)存資源。

目前，MLOPart 僅支持 NVIDIA B200 與 NVIDIA B300 系列產(chǎn)品。未來的 CUDA 發(fā)布版本將加入對 NVIDIA GB200 與 NVIDIA GB300 系列的支持。

靜態(tài)流式多處理器（SM）分區(qū)

作為 MPS 中現(xiàn)有的動態(tài)執(zhí)行資源供給（provisioning）的一種替代方案，靜態(tài)流式多處理器（SM）分區(qū)是針對 NVIDIA Ampere 架構(gòu)（計算能力 8.0）及更新 GPU 的一項特性，它為 MPS 客戶端提供了一種創(chuàng)建獨占 SM 分區(qū)的方法。

該模式通過使用 -S 或 --static-partitioning 標志啟動 MPS 控制守護進程來啟用，其主要目的是提供確定性的資源分配，并改善 MPS 客戶端之間的隔離性。分區(qū)的基本單位是一個「Chunk」（塊），其大小根據(jù) GPU 架構(gòu)而異 —— 例如，在 Hopper（計算能力 9.0）及更新的獨立 GPU 上，一個 Chunk 包含 8 個 SM。

cuBLAS 中的雙精度和單精度模擬

雖然嚴格來說這不屬于 CUDA 13.1 的更新，但 NVIDIA CUDA Toolkit 13.0 中的 cuBLAS 更新引入了新的 API 和實現(xiàn)，旨在提升雙精度（FP64）矩陣乘法（matmul）的性能。

這是通過在 NVIDIA GB200 NVL72 和 NVIDIA RTX PRO 6000 Blackwell Server Edition 等 GPU 架構(gòu)的 Tensor Core 上進行浮點（FP）模擬來實現(xiàn)的。

開發(fā)者工具

開發(fā)者工具是 CUDA 平臺的重要組成部分。此次發(fā)布帶來了多項創(chuàng)新和功能增強，包括：

CUDA Tile 核函數(shù)性能分析工具

• 在摘要頁新增「Result Type」（結(jié)果類型）列，用于區(qū)分 Tile 核函數(shù)與 SIMT 核函數(shù)。
• 詳情頁新增「Tile Statistics」（Tile 統(tǒng)計）部分，總結(jié) Tile 維度和重要管線（pipeline）的利用率。
• 源碼頁支持將指標映射到高層級的 cuTile 核函數(shù)源碼。

Nsight Compute 分析，重點展示了分析輸出中的 Tile Statistics 部分

此次發(fā)布的 Nsight Compute 還增加了對設備端啟動的圖（device-launched graphs）中 CUDA 圖節(jié)點的分析支持，并改進了源碼頁導航，為編譯器生成和用戶生成的標簽提供了可點擊的鏈接。

編譯時修補

NVIDIA Compute Sanitizer 2025.4 通過 -fdevice-sanitize=memcheck 編譯器標志，增加了對 NVIDIA CUDA 編譯器（NVCC）編譯時修補（patching）的支持。這種修補增強了內(nèi)存錯誤檢測能力，并提升了 Compute Sanitizer 的性能。

編譯時插樁（instrumentation）可將錯誤檢測直接集成到 NVCC 中，從而實現(xiàn)更快的運行速度，并通過高級的基址 - 邊界分析（base-and-bounds analysis）捕捉更隱蔽的內(nèi)存問題（如相鄰分配間的非法訪問）。這意味著開發(fā)者可以在不犧牲速度的情況下調(diào)試內(nèi)存問題，運行更多測試并保持生產(chǎn)力。目前，該功能僅支持 memcheck 工具。

要使用此新功能，請使用如下 NVCC 標志編譯代碼：

nvcc -fdevice-sanitize=memcheck -o myapp myapp.cu

然后使用 memcheck 工具運行你的應用：

compute-sanitizer --tool memcheck myapp

NVIDIA Nsight Systems

NVIDIA Nsight Systems 2025.6.1 與 CUDA Toolkit 13.1 同步發(fā)布，帶來了多項新的追蹤功能：

• 系統(tǒng)級 CUDA 追蹤：--cuda-trace-scope 可開啟跨進程樹或整個系統(tǒng)的追蹤。
• CUDA 主機函數(shù)追蹤：增加了對 CUDA Graph 主機函數(shù)節(jié)點和 cudaLaunchHostFunc () 的追蹤支持，這些函數(shù)在主機上執(zhí)行并會阻塞流（stream）。
• CUDA 硬件追蹤：在支持的情況下，基于硬件的追蹤現(xiàn)在成為默認模式；使用 --trace=cuda-sw 可恢復為軟件模式。
• Green Context 時間軸行現(xiàn)在會在工具提示中顯示 SM 分配情況，幫助用戶理解 GPU 資源利用率。

數(shù)學庫

核心 CUDA 工具包數(shù)學庫的新功能包括：

• NVIDIA cuBLAS：一項全新的實驗性 API，支持 Blackwell GPU 的分組 GEMM 功能，并兼容 FP8 和 BF16/FP16 數(shù)據(jù)類型。針對上述數(shù)據(jù)類型，支持 CUDA 圖的分組 GEMM 提供了一種無需主機同步的實現(xiàn)方式，其設備端形狀可實現(xiàn)最高 4 倍的加速，優(yōu)于 MoE 用例中的多流 GEMM 實現(xiàn)。
• NVIDIA cuSPARSE：一種新的稀疏矩陣向量乘法 (SpMVOp) API，與 CsrMV API 相比性能有所提升。該 API 支持 CSR 格式、32 位索引、雙精度以及用戶自定義的后綴。
• NVIDIA cuFFT：一套名為 cuFFT 設備 API 的全新 API，提供主機函數(shù)，用于在 C++ 頭文件中查詢或生成設備功能代碼和數(shù)據(jù)庫元數(shù)據(jù)。該 API 專為 cuFFTDx 庫設計，可通過查詢 cuFFT 來生成 cuFFTDx 代碼塊，這些代碼塊可以與 cuFFTDx 應用程序鏈接，從而提升性能。

針對新的 Blackwell 架構(gòu)，現(xiàn)已推出性能更新。用戶可選擇關(guān)鍵 API 進行更新，并查看性能更新詳情。

cuBLAS Blackwell 性能

CUDA Toolkit 12.9 在 NVIDIA Blackwell 平臺上引入了塊縮放的 FP4 和 FP8 矩陣乘法。CUDA 13.1 增加了對這些數(shù)據(jù)類型和 BF16 的性能支持。圖 2 顯示了在 NVIDIA Blackwell 和 Hopper 平臺上的加速比。

cuSOLVER Blackwell 性能

CUDA 13.1 繼續(xù)優(yōu)化用于特征分解的批處理 SYEVD 與 GEEV API，并帶來了顯著的性能增強。

其中，批處理 SYEV（cusolverDnXsyevBatched） 是 cuSOLVER 中 SYEV 例程的統(tǒng)一批處理版本，用于計算對稱／Hermitian 矩陣的特征值與特征向量，非常適合對大量小矩陣進行并行求解的場景。

圖 3 展示了在批大小為 5,000（矩陣行數(shù) 24–256）的測試結(jié)果。與 NVIDIA L40S 相比，NVIDIA Blackwell RTX Pro 6000 Server Edition 實現(xiàn)了約 2 倍的加速，這與預期的內(nèi)存帶寬提升相吻合。

對于復數(shù)單精度和實數(shù)單精度兩類矩陣，當行數(shù)N = 5時，加速比約為1.5×，并隨著行數(shù)增大逐漸提升，在N = 250 時達到 2.0×。

圖 4 顯示了 cusolverDnXgeev (GEEV) 的性能加速比，該函數(shù)用于計算一般（非對稱）稠密矩陣的特征值和特征向量。GEEV 是一種混合 CPU/GPU 算法。單個 CPU 線程負責在 QR 算法中執(zhí)行高效的早期降階處理，而 GPU 則處理其余部分。圖中顯示了矩陣大小從 1,024 到 32,768 的相對性能加速比。

當矩陣行數(shù)n = 5000時，加速比約為1.0，并隨著矩陣規(guī)模增大逐漸提升，在n = 30000 時達到約 1.7。

NVIDIA CUDA 核心計算庫

NVIDIA CUDA Core 計算庫 (CCCL) 為 CUB 帶來了多項創(chuàng)新和增強功能。

確定性浮點運算簡化

由于浮點加法不具備結(jié)合律，cub::DeviceReduce 歷史上只能保證在同一 GPU 上每次運行得到位上完全相同的結(jié)果。這被實現(xiàn)為一個兩遍算法。

作為 CUDA 13.1 的一部分， NVIDIA CCCL 3.1 提供了兩個額外的浮點確定性選項，您可以根據(jù)這些選項在確定性和性能之間進行權(quán)衡。

• 不保證：使用原子操作進行單次歸約。這不能保證提供位上完全相同的結(jié)果。
• GPU 間：基于 Kate Clark 在 NVIDIA GTC 2024 大會上演講中可復現(xiàn)的降維結(jié)果。結(jié)果始終逐位相同。

可以通過標志位設置確定性選項，如下面的代碼所示。

更便捷的單相 CUB API

幾乎所有 CUB 算法都需要臨時存儲空間作為中間暫存空間。過去，用戶必須通過兩階段調(diào)用模式來查詢和分配必要的臨時存儲空間，如果兩次調(diào)用之間傳遞的參數(shù)不一致，這種模式既繁瑣又容易出錯。

CCCL 3.1 為一些接受內(nèi)存資源的 CUB 算法添加了新的重載，從而用戶可以跳過臨時存儲查詢 / 分配 / 釋放模式。

• CUDA Tile 資源鏈接：https://developer.nvidia.com/cuda/tile
• CUDA Toolkit 13.1 下載地址：https://developer.nvidia.com/cuda-downloads

https://developer.nvidia.com/blog/focus-on-your-algorithm-nvidia-cuda-tile-handles-the-hardware

https://developer.nvidia.com/blog/nvidia-cuda-13-1-powers-next-gen-gpu-programming-with-nvidia-cuda-tile-and-performance-gains

https://x.com/NVIDIAAIDev/status/1996976702732620271

https://developer.nvidia.com/blog/simplify-gpu-programming-with-nvidia-cuda-tile-in-python