企业官网建设流程全解析

CUDA Samples 实战：用 batchCUBLAS 批处理 API 提升 CUBLAS 性能

【免费下载链接】cuda-samplesSamples for CUDA Developers which demonstrates features in CUDA Toolkit项目地址: https://gitcode.com/GitHub_Trending/cu/cuda-samples

导读

本篇文章围绕 NVIDIA CUDA Samples 仓库中的 batchCUBLAS 示例（cpp/4_CUDA_Libraries/batchCUBLAS）展开，深入讲解如何利用 CUBLAS 的批处理（batched）API 与 CUDA Stream 技术，将多个独立的矩阵乘（GEMM）调用合并为一次批量提交，从而显著提升整体吞吐。阅读本文后，你将掌握：batchCUBLAS 的三种执行模式（普通循环、多 Stream、批处理 API）的实现差异、示例的命令行参数与运行方式、以及如何从源码层面理解其性能对比与正确性校验逻辑。

示例概述与关键概念

示例做了什么

根据 README.md 的描述，这是一个用于演示通过批处理 CUBLAS API 调用来提升整体性能的 CUDA 示例。其核心思想是：当需要执行大量相同形状的矩阵乘法时，如果逐个发起cublasSgemm/cublasDgemm调用，会产生大量内核启动开销；而使用cublasSgemmBatched/cublasDgemmBatched一次性提交一批矩阵运算，可以大幅削减启动成本，提高 GPU 利用率。

注意：batchCUBLAS 侧重的是"调用方式"层面的批处理（Batched API + Stream），与使用 cuBLASLt 的cublasLtMatmul或其他高级批处理抽象不同，它的学习价值在于理解 Batched API 的经典用法与性能收益。

关键概念

• Linear Algebra（线性代数）：示例的计算核心是 GEMM（通用矩阵乘），即C = alpha * op(A) * op(B) + beta * C。
• CUBLAS Library：NVIDIA 提供的 BLAS 库的 CUDA 实现，批处理函数是自 CUDA 4.1 起引入的特性（源码中以#if CUDART_VERSION >= 4010做了版本守护）。

支持的平台与架构

依据 README.md 中的官方声明：

而在实际构建层面，CMakeLists.txt 中预设的编译目标架构为75 80 86 87 89 90 100 110 120（覆盖 Turing 至 Blackwell 世代），CMake 会按本机 GPU 能力自动挑选合适的架构进行编译。

涉及的 CUDA API

README 中列出了本示例涉及的两类 API：

CUDA Runtime API（核心）

• cudaMalloc/cudaFree：设备内存的分配与释放；
• cudaMemcpy：主机与设备间的数据拷贝（在批处理模式下还需把指针数组拷贝到设备端，见下文源码分析）；
• cudaStreamCreate：为"多 Stream 模式"创建并发流；
• cudaDeviceSynchronize：等待内核执行完成并获取计时；
• cudaGetDevice/cudaGetDeviceProperties：查询当前设备及属性（用于判断设备计算能力是否支持双精度）；
• cudaGetErrorString/cudaGetLastError：错误信息获取与异步错误捕获。

CUDA Driver API 相关辅助函数

README 提到cuRand、cuEqual，这两者实际上定义在 batchCUBLAS.h 中：

• cuRand()是一个基于 George Marsaglia KISS 算法的快速伪随机数生成器（batchCUBLAS.h），用于在测试中随机生成alpha/beta等标量；
• cuEqual()是一个跨__host__/__device__的类型安全比较模板（batchCUBLAS.h），用于判断beta是否为零以决定是否需要对C矩阵预填充。

依赖与前置条件

README 明确指出，构建/运行本示例需要：

1. CUBLAS：即 CUDA Toolkit 自带的 cuBLAS 库。仓库顶层 README.md 中列有 CUBLAS 相关依赖说明，本示例只需在 CMake 中链接CUDA::cublas（见 CMakeLists.txt）；
2. CUDA Toolkit：从 NVIDIA 官方渠道下载并安装与你平台对应的 CUDA Toolkit；
3. 确保上述依赖已正确安装后再进行编译。

在源码层面，本示例还引入了仓库公共头文件 helper_cuda.h（用于findCudaDevice等设备选择工具），编译时通过include_directories(../../../Common)指向仓库的 Common 目录。

构建与运行指南

使用 CMake 构建

与仓库其他示例一致，batchCUBLAS 支持通过仓库顶层 CMake 工程统一构建，也可以单独进入示例目录构建。以 Linux 为例（完整流程参考仓库顶层 README.md）：

cd cuda-samples mkdir build && cd build cmake .. # 配置工程，会按 add_subdirectory(batchCUBLAS) 纳入本示例 make -j$(nproc) # 并行编译

Windows 下可使用cmake .. -G "Visual Studio 16 2019" -A x64生成 VS 工程后编译。

构建成功后，可执行文件位于build/4_CUDA_Libraries/batchCUBLAS/batchCUBLAS（具体输出路径随 CMake 配置而定），程序运行时需保证 GPU 驱动与 CUDA 运行时环境正常。

命令行参数

main()中通过processArgs()（batchCUBLAS.cpp）解析参数，完整用法如下：

batchcublas [-mSIZE_M] [-nSIZE_N] [-kSIZE_K] [-NSIZE_NUM_ITERATIONS] [-qatest] [-noprompt]

各参数含义（对照源码）：

矩阵尺寸的默认值定义在 batchCUBLAS.cpp，未显式指定时取128 x 128 x 128；-N默认 10 次乘法的设置在 batchCUBLAS.cpp。

例如，执行 64 次 256×256×256 的批量 GEMM 对比测试：

./batchCUBLAS -m256 -n256 -k256 -N64

三种执行模式的源码级剖析

main()（batchCUBLAS.cpp）是整个示例的主控逻辑，它依次运行：

1. 单内核预热：分别以sgemm（float）和dgemm（double）各执行一次单次 GEMM，并校验结果正确性；
2. 三轮性能对比（tmRegular→tmStream→tmBatched）：在相同矩阵尺寸与相同批次数N下，分别用三种方式执行同样的 GEMM 序列，并打印各自的耗时与 GFLOPS。

三种模式由枚举enum testMethod { tmRegular, tmStream, tmBatched }（batchCUBLAS.cpp）区分。

模式一：普通循环（tmRegular）

这是最直观的做法：在for循环中逐个调用cublasXgemm（模板化包装，float 走cublasSgemm、double 走cublasDgemm，见 batchCUBLAS.cpp），每次调用前通过cublasSetStream(handle, streamArray[i])将句柄绑定到默认流（streamArray[i] = 0）。

for (int i = 0; i < opts.N; i++) { cublasSetStream(handle, streamArray[i]); status1 = cublasXgemm(handle, params.transa, params.transb, params.m, params.n, params.k, &params.alpha, devPtrA[i], rowsA, devPtrB[i], rowsB, &params.beta, devPtrC[i], rowsC); // ...错误检查 }

这段代码位于 batchCUBLAS.cpp。由于所有调用都在默认流上串行执行，N次内核依次排队，启动开销随N线性增长，是三种模式中的性能基线。

模式二：多 Stream（tmStream）

模式二在进入测试前为每个批次创建独立 CUDA Stream（batchCUBLAS.cpp）：

if (opts.test_method == tmStream) { cudaError_t cudaErr = cudaStreamCreate(&streamArray[i]); ... }

随后同样逐个调用cublasXgemm，但每次调用前把句柄切换到对应的streamArray[i]。由于不同流上的内核可以并行执行（前提是流之间无数据依赖且 GPU 资源充足），整体执行时间可以比模式一明显缩短。不过每个内核仍是独立启动，N次启动的 CPU 侧调度开销依然存在。

模式三：批处理 API（tmBatched）—— 示例的核心

模式三使用cublasSgemmBatched/cublasDgemmBatched（模板包装见 batchCUBLAS.cpp），一次调用完成全部N个矩阵乘：

status1 = cublasXgemmBatched(handle, params.transa, params.transb, params.m, params.n, params.k, &params.alpha, (const T_ELEM **)devPtrA_dev, rowsA, (const T_ELEM **)devPtrB_dev, rowsB, &params.beta, devPtrC_dev, rowsC, opts.N);

这段代码位于 batchCUBLAS.cpp。批处理模式有一个关键差异：cublasXgemmBatched要求传入的Aarray/Barray/Carray是设备端指针数组，因此示例在进入测试前额外执行了三步准备工作（batchCUBLAS.cpp）：

1. cudaMalloc为指针数组本身分配设备内存（devPtrA_dev、devPtrB_dev、devPtrC_dev）；
2. 用cudaMemcpy(..., cudaMemcpyHostToDevice)把主机端各批次矩阵指针数组拷贝到设备端；
3. 之后将这组设备端指针数组一次性传给批处理 API。

由于只需一次内核启动即可处理N个矩阵乘，内核启动开销被大幅摊薄，这是三种模式中理论上性能最优的方案，也正是 README 所说"batched CUBLAS API calls to improve overall performance"的直接体现。

值得注意的是：批处理模式把每个矩阵的指针数组放在设备内存中，这就要求各矩阵的lda/ldb/ldc（leading dimension）一致、且矩阵在内存中是独立连续分配的——这是使用 Batched API 时最重要的内存布局前提。

正确性校验与性能计量

测试参数生成

每次循环通过TESTGEN(gemm)（即get_gemm_params，batchCUBLAS.cpp）随机生成一组alpha、beta与矩阵维度，其中alpha/beta从{0, -1, 1, 2, 0}等候选值中选取（cuRand() % NBR_ALPHAS），同时固定transa = transb = CUBLAS_OP_N。

结果正确性检查

示例定义了两种容差指标（batchCUBLAS.cpp）：

#define CUBLAS_SGEMM_MAX_ULP_ERR (.3) // float 版最大 ULP 误差 #define CUBLAS_DGEMM_MAX_ULP_ERR (1.e-3) // double 版最大 ULP 误差 #define CUBLAS_SGEMM_MAX_RELATIVE_ERR (6.e-6) // float 版最大相对误差 #define CUBLAS_DGEMM_MAX_RELATIVE_ERR (0.0) // double 版最大相对误差

此外还通过边界填充技巧加强校验强度（batchCUBLAS.cpp）：

• 用memset(A, 0xFF, ...)先把矩阵填充为 NaN 位模式，再调用fillupMatrixDebug/fillupMatrix覆盖有效区域，从而验证 lda padding 区域不会被内核错误访问；
• 当beta == 0时，将C填充为 NaN，确保 GEMM 在beta=0语义下不会读取C的旧值（这是 BLAS 规范的行为）。

性能计量

计时使用跨平台second()函数（Windows 走QueryPerformanceCounter，Linux/QNX/Apple 走gettimeofday，见 batchCUBLAS.h）。每次测量前后分别调用cudaDeviceSynchronize()保证内核全部完成，然后输出：

^^^^ elapsed = %10.8f sec GFLOPS=%g

GFLOPS 的计算公式为opts.N * (2 * m * n * k) / elapsed（batchCUBLAS.cpp），其中系数 2 对应一次矩阵乘的乘加次数。通过对比三种模式下输出的elapsed与GFLOPS，即可直观看到批处理带来的性能提升。

双精度支持检测

由于部分低端 GPU 不支持双精度运算，示例通过getDeviceVersion()（batchCUBLAS.cpp）读取设备计算能力，当major*100 + minor*10 < 130（即计算能力低于 1.3，对应DEV_VER_DBL_SUPPORT）时跳过dgemm测试并输出WAIVED提示（batchCUBLAS.cpp）。

运行结果解读

以默认参数运行，程序输出的大致结构如下（具体数值随 GPU 与驱动而异）：

batchCUBLAS Starting... ==== Running single kernels ==== Testing sgemm ... @@@@ sgemm test OK Testing dgemm ... @@@@ dgemm test OK ==== Running N=10 without streams ==== Testing sgemm ^^^^ elapsed = ... sec GFLOPS=... ... ==== Running N=10 with streams ==== ... ==== Running N=10 batched ==== ... Test Summary 0 error(s)

阅读要点：

• Running single kernels阶段验证 sgemm/dgemm 的正确性，若输出FAIL或出现!!!! GPU program execution error，需检查 GPU 驱动、显存大小与矩阵维度是否合理；
• 三段Running N=...输出分别对应普通循环、多 Stream、批处理三种模式，GFLOPS越高代表吞吐越好；
• 最后Test Summary汇总错误数，0 error(s)表示全部通过。注意：当-N较小时，多 Stream 与批处理的优势可能不明显；增大-N（例如-N1000）更能体现批处理带来的启动开销摊薄收益。

小结

batchCUBLAS 示例用最简练的代码完整呈现了"批量矩阵乘"的三种实现路径：普通循环、多 Stream 并行、Batched API。从 batchCUBLAS.cpp 与 batchCUBLAS.h 的源码可以看到，Batched API 的核心在于把"设备端指针数组"一次提交给cublasXgemmBatched，从而用一次内核启动替代N次启动。对需要频繁执行同形状小规模矩阵乘的场景（如批量求解、批处理推理中的逐样本 GEMM），这是兼顾代码简洁与性能的推荐写法；而若矩阵尺寸差异较大，则需评估 Batched API 的固定 leading dimension 约束是否适用，必要时可转向 cuBLASLt 等更灵活的接口。结合本示例的-m/-n/-k/-N参数，读者可以在自己的 GPU 上快速复现三种模式的性能差异，为实际项目中的批量线性代数方案选型提供第一手数据。

【免费下载链接】cuda-samplesSamples for CUDA Developers which demonstrates features in CUDA Toolkit项目地址: https://gitcode.com/GitHub_Trending/cu/cuda-samples