企业官网建设流程全解析

CUTLASS GPU 矩阵乘法 3 步跑通：CUDA C++ 模板与 Python DSL 完整指南

【免费下载链接】cutlassCUDA Templates and Python DSLs for High-Performance Linear Algebra项目地址: https://gitcode.com/GitHub_Trending/cu/cutlass

CUTLASS 是 NVIDIA 的 CUDA C++ 模板库，外加一套 Python DSL，把 GEMM 矩阵乘、卷积和各种低精度计算压成可组合的模板。它帮做深度学习、科学计算的工程师省掉手写 kernel 的繁复细节，直接拿到接近峰值的 GPU 性能。

先说痛点：为什么手写 CUDA kernel 总差点意思

手写 GEMM，细节全靠自己

矩阵乘看着简单，跑得快难。共享内存分块、bank conflict、流水线深度，每一步都直接决定性能。更麻烦的是，Volta 到 Blackwell，每代 GPU 的张量核心指令都不同，老代码到新架构就得重写。

图：GEMM 中的分块策略，A、B 矩阵按块切分，逐块计算累加成 C

分层模板替代逐行手写

CUTLASS 把计算拆成设备、内核、CTA、warp、线程、指令六层，每层都有现成实现。你只在需要定制的层级动手，其余交给库。00_basic_gemm 这个示例里，你写的只有布局和块大小，矩阵搬运、warp 分片、指令发射全由模板完成。

多精度支持一套模板全覆盖

FP32、TF32、FP16、BF16 到 FP8、INT4，每换一种精度，naive 写法就得重做一遍数据排布。CUTLASS 用同一个模板体系覆盖这些精度：换精度就是换几个模板参数，对应的张量核心指令和寄存器组织由库处理。3.5.1 版本起还支持 Hopper、Ada、Blackwell 上的一批新格式。

图：CUTLASS 3.5.1 在 H100 SXM5 上各精度、各矩阵尺寸的 GEMM 性能基准

算子组合不用自己焊 kernel

GEMM 之后接 bias、ReLU，再接逐元素运算，纯手写往往要写三个 kernel 加两次显存往返。CUTLASS 的 epilogue 和融合接口让你在同一 kernel 里把它们串起来。13_two_tensor_op_fusion 演示的就是这类多算子融合。

三步跑通第一个 GEMM

环境就一条：装好 CUDA

克隆仓库即可：

git clone https://gitcode.com/GitHub_Trending/cu/cutlass

编译用 nvcc，CMake 会帮你把 include/ 头文件路径接好。

20 行代码调起 GEMM

C++ 侧的核心就三段：声明类型与布局、填参数、发起调用。

using Gemm = cutlass::gemm::device::Gemm< float, cutlass::layout::RowMajor, float, cutlass::layout::ColumnMajor, float, cutlass::layout::RowMajor, float>; // 累加器类型 Gemm::Arguments cfg{ {M, N, K}, // 问题尺寸 A, lda, B, ldb, C, ldc, D, ldd, {alpha, beta} // 缩放系数 }; Gemm op; op(cfg); // 检查返回的 Status

完整可编译版本在 00_basic_gemm。想少写代码，也可以走 Python 绑定，见 python/。

底层原理：六层积木，各管一段

图：CUTLASS 的分层组织，从设备级调度到指令级硬件模板

读法很简单：最上面设备级负责在主机侧启动 kernel；内核级是一整个 GEMM kernel 的骨架；CTA 级管块内所有线程的集体协作，GEMM 主循环就住在这层；warp 级对应张量核心指令的同步执行；线程级和指令级则贴近 SIMT 与硬件指令。

你日常改的是前几层，指令级基本不用碰。以 basic_gemm 为例：库把矩阵切成块，从全局内存搬进共享内存，再分发到每个 warp 的寄存器，最后用硬件矩阵乘指令累加。整条流水线你一行没写。

进阶实战：两个真实场景

Blackwell 上的低延迟 GQA

LLM 解码阶段的 GQA（Grouped Query Attention）batch 小，延迟敏感。93_blackwell_low_latency_gqa 示例把 CTA 的切分方式重新组织，让每次 GEMV 型计算的占用率更合理；累加结果暂存在寄存器里的"CTA 邮箱"中，配合 TMA 异步搬运，计算和搬数重叠进行，从而压低解码延迟。

图：该示例的 CTA 组织方式，展示了 GQA 计算如何切分与映射

INT4 卷积前向：视觉推理的常用组合

图像和视频推理里，INT4 权重量化已经普及。CUTLASS 把卷积改写成隐式 GEMM：激活和滤波器先排成矩阵形状，再交给低精度张量核心算。16_ampere_tensorop_conv2dfprop 给了完整参考。

图：INT4 精度下的卷积前向传播，激活与滤波器张量的处理流程

学习路径：仓库里都放好了

• C++ API 文档：docs/，每个类、头文件都有页面
• 示例代码：examples/，95 个以上按编号排列的场景，从 basic_gemm 到 Blackwell
• Profiling 工具：指南在 media/docs/cpp/profiler.md，命令行直接对比配置性能
• Python 接口：python/ 下有 cutlass_library 绑定和 CuTe 的 Python DSL
• CuTe 教程：examples/cute/tutorial，官方 CuTe GEMM 教程系列

建议先 clone 仓库跑通 basic_gemm，再从上面挑一个和你的业务最近的例子精读。

【免费下载链接】cutlassCUDA Templates and Python DSLs for High-Performance Linear Algebra项目地址: https://gitcode.com/GitHub_Trending/cu/cutlass