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

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 的批处理batchedAPI 与 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 LibraryNVIDIA 提供的 BLAS 库的 CUDA 实现批处理函数是自 CUDA 4.1 起引入的特性源码中以#if CUDART_VERSION 4010做了版本守护。支持的平台与架构依据 README.md 中的官方声明维度支持范围SM 架构SM 5.0、5.2、5.3、6.0、6.1、7.0、7.2、7.5、8.0、8.6、8.7、8.9、9.0操作系统Linux、WindowsCPU 架构x86_64、armv7l而在实际构建层面CMakeLists.txt 中预设的编译目标架构为75 80 86 87 89 90 100 110 120覆盖 Turing 至 Blackwell 世代CMake 会按本机 GPU 能力自动挑选合适的架构进行编译。涉及的 CUDA APIREADME 中列出了本示例涉及的两类 APICUDA 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 明确指出构建/运行本示例需要CUBLAS即 CUDA Toolkit 自带的 cuBLAS 库。仓库顶层 README.md 中列有 CUBLAS 相关依赖说明本示例只需在 CMake 中链接CUDA::cublas见 CMakeLists.txtCUDA Toolkit从 NVIDIA 官方渠道下载并安装与你平台对应的 CUDA Toolkit确保上述依赖已正确安装后再进行编译。在源码层面本示例还引入了仓库公共头文件 helper_cuda.h用于findCudaDevice等设备选择工具编译时通过include_directories(../../../Common)指向仓库的 Common 目录。构建与运行指南使用 CMake 构建与仓库其他示例一致batchCUBLAS 支持通过仓库顶层 CMake 工程统一构建也可以单独进入示例目录构建。以 Linux 为例完整流程参考仓库顶层 README.mdcd 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]各参数含义对照源码参数说明默认值源码-m矩阵 A 的行数M128BENCH_MATRIX_M-n矩阵 B 的列数N128BENCH_MATRIX_N-k矩阵 A 的列数 / B 的行数K128BENCH_MATRIX_K-N批处理中矩阵乘的次数number of multiplications10-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是整个示例的主控逻辑它依次运行单内核预热分别以sgemmfloat和dgemmdouble各执行一次单次 GEMM并校验结果正确性三轮性能对比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线性增长是三种模式中的性能基线。模式二多 StreamtmStream模式二在进入测试前为每个批次创建独立 CUDA StreambatchCUBLAS.cppif (opts.test_method tmStream) { cudaError_t cudaErr cudaStreamCreate(streamArray[i]); ... }随后同样逐个调用cublasXgemm但每次调用前把句柄切换到对应的streamArray[i]。由于不同流上的内核可以并行执行前提是流之间无数据依赖且 GPU 资源充足整体执行时间可以比模式一明显缩短。不过每个内核仍是独立启动N次启动的 CPU 侧调度开销依然存在。模式三批处理 APItmBatched—— 示例的核心模式三使用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.cppcudaMalloc为指针数组本身分配设备内存devPtrA_dev、devPtrB_dev、devPtrC_dev用cudaMemcpy(..., cudaMemcpyHostToDevice)把主机端各批次矩阵指针数组拷贝到设备端之后将这组设备端指针数组一次性传给批处理 API。由于只需一次内核启动即可处理N个矩阵乘内核启动开销被大幅摊薄这是三种模式中理论上性能最优的方案也正是 README 所说batched CUBLAS API calls to improve overall performance的直接体现。值得注意的是批处理模式把每个矩阵的指针数组放在设备内存中这就要求各矩阵的lda/ldb/ldcleading dimension一致、且矩阵在内存中是独立连续分配的——这是使用 Batched API 时最重要的内存布局前提。正确性校验与性能计量测试参数生成每次循环通过TESTGEN(gemm)即get_gemm_paramsbatchCUBLAS.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 在beta0语义下不会读取C的旧值这是 BLAS 规范的行为。性能计量计时使用跨平台second()函数Windows 走QueryPerformanceCounterLinux/QNX/Apple 走gettimeofday见 batchCUBLAS.h。每次测量前后分别调用cudaDeviceSynchronize()保证内核全部完成然后输出^^^^ elapsed %10.8f sec GFLOPS%gGFLOPS 的计算公式为opts.N * (2 * m * n * k) / elapsedbatchCUBLAS.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 N10 without streams Testing sgemm ^^^^ elapsed ... sec GFLOPS... ... Running N10 with streams ... Running N10 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创作声明：本文部分内容由AI辅助生成（AIGC），仅供参考