cuda-samples 中的 conjugateGradientCudaGraphs用 CUDA Graph 捕获 CUBLAS/CUSPARSE 的共轭梯度求解器【免费下载链接】cuda-samplesSamples for CUDA Developers which demonstrates features in CUDA Toolkit项目地址: https://gitcode.com/GitHub_Trending/cu/cuda-samples导读conjugateGradientCudaGraphs 是 CUDA Samplescuda-samples仓库中一个以共轭梯度Conjugate Gradient, CG迭代求解稀疏线性方程组为载体的 CUDA Graph 演示样例。它把 CUBLAS 的向量运算与 CUSPARSE 的稀疏矩阵向量乘SpMV完整地捕获进一张 CUDA Graph从而以一次性提交、反复回放的方式执行每一轮迭代展示了 Graph API 降低内核启动开销的典型用法。读完本文你将掌握该样例的数据组织方式CSR 三对角矩阵、CG 迭代的库函数调用序列、CUDA Graph 的捕获—实例化—启动—销毁全流程以及如何编译、运行并验证其结果。示例概览从普通 CG 到 Graph 化 CG根据该示例自带的 README.md 描述本样例在 GPU 上实现了一个共轭梯度求解器其所有 CUBLAS 与 CUSPARSE 库调用都被 CUDA Graph API 捕获并调用a conjugate gradient solver on GPU using CUBLAS and CUSPARSE library calls captured and called using CUDA Graph APIs。关键概念Key Concepts线性代数Linear Algebra、CUBLAS 库、CUSPARSE 库。支持的操作系统Linux、Windows。支持的 CPU 架构x86_64、armv7l。构建/运行依赖CUBLAS、CUSPARSE均随 CUDA Toolkit 一并提供安装对应平台的 CUDA Toolkit 即可满足前置条件。在仓库的 4_CUDA_Libraries 目录下它与普通版conjugateGradient、conjugateGradientPrecond、conjugateGradientUM、conjugateGradientMultiDeviceCG等构成一个完整的 CG 求解器家族便于读者横向对比非 Graph 版与Graph 版在实现上的差异。算法与 Graph 化的动机共轭梯度法用于求解对称正定SPD稀疏线性方程组A x b。每轮迭代由以下基本运算组成计算方向向量p的 SpMVAx A * pCUSPARSE若干点积dot与 AXPY 形式的向量更新CUBLAS标量运算本例中由两个极小的自定义 kernel 完成。当这些库调用密集且规模很小时每次 CPU 提交单个 kernel/库调用的启动开销占比会非常可观。CUDA Graph 允许把一段流stream上的完整操作序列录制成图之后每次迭代只需一次cudaGraphLaunch即可整图回放从而显著摊薄启动成本。这正是该样例把 CG 迭代体捕获成图的原因也是它与 conjugateGradient同算法、非 Graph 实现形成对照实验的价值所在。编译与运行该样例通过 CMake 构建其 CMakeLists.txt 中关键配置如下编译目标conjugateGradientCudaGraphs源文件为单文件 conjugateGradientCudaGraphs.cu链接库为CUDA::cublas与CUDA::cusparse对应 README 中声明的两项依赖默认编译架构CMAKE_CUDA_ARCHITECTURES覆盖 75/80/86/87/89/90/100/110/120即 Volta 到 Blackwell 主流代际开启CUDA_SEPARABLE_COMPILATION并启用--extended-lambda编译选项启用ENABLE_CUDA_DEBUG时追加-G配合 cuda-gdb否则默认追加-lineinfo便于性能剖析。构建方式与仓库其他样例一致在仓库根目录下创建构建目录并执行 CMake 配置、编译之后可在构建目录中运行生成的可执行文件程序运行前会自动通过findCudaDevice选择最佳可用 GPU详见源码 conjugateGradientCudaGraphs.cu。数据准备随机三对角对称矩阵的 CSR 构造程序固定求解规模为N 1048576100 万阶的三对角对称线性系统非零元个数nz (N-2)*3 4。矩阵以CSRCompressed Sparse Row格式存储包含三个数组I行指针长度N1J列索引长度nzval非零值长度nz。由 genTridiag 生成主对角线元素取rand()/RAND_MAX 10.0保证对角占优、对称正定次对角线元素取rand()/RAND_MAX并按三对角结构填充I/J索引。初始条件为右端项b源码中的rhs全 1解向量x全 0初猜残差即b本身。主机端数据通过cudaMallocHost分配页锁定内存随后经异步拷贝cudaMemcpyAsync上传到设备端的d_row/d_col/d_val。设备端随后把原始 CSR 数据包装进 CUSPARSE 通用 API 的对象cusparseCreateCsr(matA, N, N, nz, d_row, d_col, d_val, CUSPARSE_INDEX_32I, CUSPARSE_INDEX_32I, CUSPARSE_INDEX_BASE_ZERO, CUDA_R_32F)创建稀疏矩阵描述符索引类型 32 位整数、基址从 0 开始、单精度浮点cusparseCreateDnVec为x、p、Ax三个向量创建稠密向量描述符cusparseSpMV_bufferSize查询 SpMV 所需工作区大小并cudaMalloc分配。此外initVectors 与cudaOccupancyMaxPotentialBlockSize配合自动计算最合适的网格/线程块规模后并行初始化设备端残差向量r与解向量x。迭代体中的库调用序列非 Graph 版在正式进入 CUDA Graph 之前程序先用普通流执行了**第一轮迭代k1**并打印残差其调用序列完整体现了 CG 的核心运算见 conjugateGradientCudaGraphs.cu步骤调用作用1cusparseSpMVCUSPARSE_OPERATION_NON_TRANSPOSE、CUSPARSE_SPMV_ALG_DEFAULT计算Ax A * x2cublasSaxpy(N, -1, d_Ax, d_r)残差r b - A*x此处即r -Ax3cublasSdot(N, r, r, d_r1)r1 r·r残差平方范数4cublasScopy(N, r, p)初始化搜索方向p r5cusparseSpMVAx A * p6cublasSdot(N, p, Ax, d_dot)dot p·Ax7kernelr1_div_x标量a r1 / dot步长 α8cublasSaxpy(N, a, p, x)更新解x a*p9kernela_minus标量na -a10cublasSaxpy(N, na, Ax, r)更新残差r - a*Ax11cudaMemcpyAsync(d_r0, d_r1)保存上一轮r1供下一轮计算 β12cublasSdot(N, r, r, d_r1)新残差平方范数13cudaMemcpyAsync(r1, d_r1)cudaStreamSynchronize回读残差用于收敛判断与打印值得注意的细节为减少不必要的同步cublasSetPointerMode被设置为CUBLAS_POINTER_MODE_DEVICE使点积结果直接写回设备端标量d_r1、d_dot而 AXPY 的标量alpha/beta等仍以主机值传入期间通过CUBLAS_POINTER_MODE_HOST与DEVICE的切换来保证正确性。CUDA Graph 捕获把整个迭代体录制成图从第二轮迭代k≥2开始若启用 Graph 路径程序将整个迭代体捕获进一张 CUDA GraphconjugateGradientCudaGraphs.cu流程如下创建专用流streamForGraph并确保 CUBLAS/CUSPARSE 句柄绑定到捕获流stream1cudaStreamBeginCapture(stream1, cudaStreamCaptureModeGlobal)以全局捕获模式开始录制在stream1上依次提交kernelr1_div_x计算 β、cublasSscalp * β、cublasSaxpyp r、cusparseSpMVAx A*p、cudaMemsetAsync清零d_dot、cublasSdot、r1_div_xα、cublasSaxpyx α*p、a_minus、cublasSaxpyr - α*Ax、cudaMemcpyAsync设备到设备保存 r0、cudaMemsetAsync清零d_r1、cublasSdot及最终残差回读cudaStreamEndCapture(stream1, initGraph)结束录制得到cudaGraph_t initGraphcudaGraphInstantiate(graphExec, initGraph, NULL, NULL, 0)实例化为可执行图cudaGraphExec_t graphExec。由于捕获流上的所有操作包括自定义 kernel 与 CUBLAS/CUSPARSE 内部 kernel都会被记录进图录制完成后每次迭代仅需cudaGraphLaunch(graphExec, streamForGraph); cudaStreamSynchronize(streamForGraph);即完成整轮迭代的提交与同步这正是 Graph 削减启动开销的核心收益。迭代循环与收敛判定主循环条件为while (r1 tol * tol k max_iter)conjugateGradientCudaGraphs.cu其中tol 1e-5收敛阈值程序比较的是残差平方范数r1与tol²等价于以残差范数sqrt(r1)与tol比较max_iter 10000最大迭代次数上限防止不收敛时死循环。循环体内在WITH_GRAPH宏默认为 1控制下二选一Graph 路径执行cudaGraphLaunch普通路径则逐条重放第 5 节的调用序列。每轮迭代打印iteration k, residual sqrt(r1)。循环退出后无论收敛还是达到上限回读解向量x并最终以exit(k max_iter ? 0 : 1)返回进程退出码——收敛则 0超限则 1方便脚本化校验。结果验证CPU 端残差检查收敛后程序在 CPU 端独立复算一遍稀疏矩阵向量乘rsum Σ val[j] * x[J[j]]并与右端项rhs[i]比较统计最大绝对误差errconjugateGradientCudaGraphs.cu最终打印Test Summary: Error amount %f。由于矩阵是三对角对称正定结构CG 在浮点精度内应快速收敛该误差即为对 GPU 求解正确性的端到端验证。涉及的 CUDA API 清单与资源管理README 列出的 CUDA Runtime API 全部可在源码中找到对应调用覆盖 Graph 全生命周期与流管理Graph 生命周期cudaStreamBeginCapture、cudaStreamEndCapture、cudaGraphInstantiate、cudaGraphLaunch、cudaGraphExecDestroy、cudaGraphDestroy流管理cudaStreamCreate、cudaStreamDestroy、cudaStreamSynchronize内存管理cudaMalloc、cudaFree、cudaMallocHost、cudaFreeHost、cudaMemcpyAsync含 H2D、D2D、D2H 三种方向、cudaMemsetAsync设备查询与内核配置cudaGetDeviceProperties、cudaOccupancyMaxPotentialBlockSize。程序退出前严格逆序释放资源先销毁 Graph 执行体与图对象、销毁流再cusparseDestroy/cublasDestroy销毁库句柄随后cusparseDestroySpMat/cusparseDestroyDnVec销毁描述符最后释放全部主机端页锁定内存与设备端显存conjugateGradientCudaGraphs.cu可作为 CUDA Graph 样例中资源管理范式的参考。小结conjugateGradientCudaGraphs 是一个麻雀虽小、五脏俱全的 CUDA Graph 教学样例它以 100 万阶三对角稀疏线性系统的 CG 求解为场景完整串联了 CUSPARSE 的 CSR 通用 APIcusparseSpMV、CUBLAS 的 BLAS-1 运算cublasSaxpy/cublasSdot/cublasSscal/cublasScopy以及 CUDA Graph 的捕获、实例化、回放与销毁全流程并提供了与非 Graph 路径的编译期开关WITH_GRAPH和 CPU 端结果校验是理解如何把多库混合的迭代算法 Graph 化以降低启动开销的理想起点。【免费下载链接】cuda-samplesSamples for CUDA Developers which demonstrates features in CUDA Toolkit项目地址: https://gitcode.com/GitHub_Trending/cu/cuda-samples创作声明：本文部分内容由AI辅助生成（AIGC），仅供参考
cuda-samples 中的 conjugateGradientCudaGraphs：用 CUDA Graph 捕获 CUBLAS/CUSPARSE 的共轭梯度求解器
cuda-samples 中的 conjugateGradientCudaGraphs用 CUDA Graph 捕获 CUBLAS/CUSPARSE 的共轭梯度求解器【免费下载链接】cuda-samplesSamples for CUDA Developers which demonstrates features in CUDA Toolkit项目地址: https://gitcode.com/GitHub_Trending/cu/cuda-samples导读conjugateGradientCudaGraphs 是 CUDA Samplescuda-samples仓库中一个以共轭梯度Conjugate Gradient, CG迭代求解稀疏线性方程组为载体的 CUDA Graph 演示样例。它把 CUBLAS 的向量运算与 CUSPARSE 的稀疏矩阵向量乘SpMV完整地捕获进一张 CUDA Graph从而以一次性提交、反复回放的方式执行每一轮迭代展示了 Graph API 降低内核启动开销的典型用法。读完本文你将掌握该样例的数据组织方式CSR 三对角矩阵、CG 迭代的库函数调用序列、CUDA Graph 的捕获—实例化—启动—销毁全流程以及如何编译、运行并验证其结果。示例概览从普通 CG 到 Graph 化 CG根据该示例自带的 README.md 描述本样例在 GPU 上实现了一个共轭梯度求解器其所有 CUBLAS 与 CUSPARSE 库调用都被 CUDA Graph API 捕获并调用a conjugate gradient solver on GPU using CUBLAS and CUSPARSE library calls captured and called using CUDA Graph APIs。关键概念Key Concepts线性代数Linear Algebra、CUBLAS 库、CUSPARSE 库。支持的操作系统Linux、Windows。支持的 CPU 架构x86_64、armv7l。构建/运行依赖CUBLAS、CUSPARSE均随 CUDA Toolkit 一并提供安装对应平台的 CUDA Toolkit 即可满足前置条件。在仓库的 4_CUDA_Libraries 目录下它与普通版conjugateGradient、conjugateGradientPrecond、conjugateGradientUM、conjugateGradientMultiDeviceCG等构成一个完整的 CG 求解器家族便于读者横向对比非 Graph 版与Graph 版在实现上的差异。算法与 Graph 化的动机共轭梯度法用于求解对称正定SPD稀疏线性方程组A x b。每轮迭代由以下基本运算组成计算方向向量p的 SpMVAx A * pCUSPARSE若干点积dot与 AXPY 形式的向量更新CUBLAS标量运算本例中由两个极小的自定义 kernel 完成。当这些库调用密集且规模很小时每次 CPU 提交单个 kernel/库调用的启动开销占比会非常可观。CUDA Graph 允许把一段流stream上的完整操作序列录制成图之后每次迭代只需一次cudaGraphLaunch即可整图回放从而显著摊薄启动成本。这正是该样例把 CG 迭代体捕获成图的原因也是它与 conjugateGradient同算法、非 Graph 实现形成对照实验的价值所在。编译与运行该样例通过 CMake 构建其 CMakeLists.txt 中关键配置如下编译目标conjugateGradientCudaGraphs源文件为单文件 conjugateGradientCudaGraphs.cu链接库为CUDA::cublas与CUDA::cusparse对应 README 中声明的两项依赖默认编译架构CMAKE_CUDA_ARCHITECTURES覆盖 75/80/86/87/89/90/100/110/120即 Volta 到 Blackwell 主流代际开启CUDA_SEPARABLE_COMPILATION并启用--extended-lambda编译选项启用ENABLE_CUDA_DEBUG时追加-G配合 cuda-gdb否则默认追加-lineinfo便于性能剖析。构建方式与仓库其他样例一致在仓库根目录下创建构建目录并执行 CMake 配置、编译之后可在构建目录中运行生成的可执行文件程序运行前会自动通过findCudaDevice选择最佳可用 GPU详见源码 conjugateGradientCudaGraphs.cu。数据准备随机三对角对称矩阵的 CSR 构造程序固定求解规模为N 1048576100 万阶的三对角对称线性系统非零元个数nz (N-2)*3 4。矩阵以CSRCompressed Sparse Row格式存储包含三个数组I行指针长度N1J列索引长度nzval非零值长度nz。由 genTridiag 生成主对角线元素取rand()/RAND_MAX 10.0保证对角占优、对称正定次对角线元素取rand()/RAND_MAX并按三对角结构填充I/J索引。初始条件为右端项b源码中的rhs全 1解向量x全 0初猜残差即b本身。主机端数据通过cudaMallocHost分配页锁定内存随后经异步拷贝cudaMemcpyAsync上传到设备端的d_row/d_col/d_val。设备端随后把原始 CSR 数据包装进 CUSPARSE 通用 API 的对象cusparseCreateCsr(matA, N, N, nz, d_row, d_col, d_val, CUSPARSE_INDEX_32I, CUSPARSE_INDEX_32I, CUSPARSE_INDEX_BASE_ZERO, CUDA_R_32F)创建稀疏矩阵描述符索引类型 32 位整数、基址从 0 开始、单精度浮点cusparseCreateDnVec为x、p、Ax三个向量创建稠密向量描述符cusparseSpMV_bufferSize查询 SpMV 所需工作区大小并cudaMalloc分配。此外initVectors 与cudaOccupancyMaxPotentialBlockSize配合自动计算最合适的网格/线程块规模后并行初始化设备端残差向量r与解向量x。迭代体中的库调用序列非 Graph 版在正式进入 CUDA Graph 之前程序先用普通流执行了**第一轮迭代k1**并打印残差其调用序列完整体现了 CG 的核心运算见 conjugateGradientCudaGraphs.cu步骤调用作用1cusparseSpMVCUSPARSE_OPERATION_NON_TRANSPOSE、CUSPARSE_SPMV_ALG_DEFAULT计算Ax A * x2cublasSaxpy(N, -1, d_Ax, d_r)残差r b - A*x此处即r -Ax3cublasSdot(N, r, r, d_r1)r1 r·r残差平方范数4cublasScopy(N, r, p)初始化搜索方向p r5cusparseSpMVAx A * p6cublasSdot(N, p, Ax, d_dot)dot p·Ax7kernelr1_div_x标量a r1 / dot步长 α8cublasSaxpy(N, a, p, x)更新解x a*p9kernela_minus标量na -a10cublasSaxpy(N, na, Ax, r)更新残差r - a*Ax11cudaMemcpyAsync(d_r0, d_r1)保存上一轮r1供下一轮计算 β12cublasSdot(N, r, r, d_r1)新残差平方范数13cudaMemcpyAsync(r1, d_r1)cudaStreamSynchronize回读残差用于收敛判断与打印值得注意的细节为减少不必要的同步cublasSetPointerMode被设置为CUBLAS_POINTER_MODE_DEVICE使点积结果直接写回设备端标量d_r1、d_dot而 AXPY 的标量alpha/beta等仍以主机值传入期间通过CUBLAS_POINTER_MODE_HOST与DEVICE的切换来保证正确性。CUDA Graph 捕获把整个迭代体录制成图从第二轮迭代k≥2开始若启用 Graph 路径程序将整个迭代体捕获进一张 CUDA GraphconjugateGradientCudaGraphs.cu流程如下创建专用流streamForGraph并确保 CUBLAS/CUSPARSE 句柄绑定到捕获流stream1cudaStreamBeginCapture(stream1, cudaStreamCaptureModeGlobal)以全局捕获模式开始录制在stream1上依次提交kernelr1_div_x计算 β、cublasSscalp * β、cublasSaxpyp r、cusparseSpMVAx A*p、cudaMemsetAsync清零d_dot、cublasSdot、r1_div_xα、cublasSaxpyx α*p、a_minus、cublasSaxpyr - α*Ax、cudaMemcpyAsync设备到设备保存 r0、cudaMemsetAsync清零d_r1、cublasSdot及最终残差回读cudaStreamEndCapture(stream1, initGraph)结束录制得到cudaGraph_t initGraphcudaGraphInstantiate(graphExec, initGraph, NULL, NULL, 0)实例化为可执行图cudaGraphExec_t graphExec。由于捕获流上的所有操作包括自定义 kernel 与 CUBLAS/CUSPARSE 内部 kernel都会被记录进图录制完成后每次迭代仅需cudaGraphLaunch(graphExec, streamForGraph); cudaStreamSynchronize(streamForGraph);即完成整轮迭代的提交与同步这正是 Graph 削减启动开销的核心收益。迭代循环与收敛判定主循环条件为while (r1 tol * tol k max_iter)conjugateGradientCudaGraphs.cu其中tol 1e-5收敛阈值程序比较的是残差平方范数r1与tol²等价于以残差范数sqrt(r1)与tol比较max_iter 10000最大迭代次数上限防止不收敛时死循环。循环体内在WITH_GRAPH宏默认为 1控制下二选一Graph 路径执行cudaGraphLaunch普通路径则逐条重放第 5 节的调用序列。每轮迭代打印iteration k, residual sqrt(r1)。循环退出后无论收敛还是达到上限回读解向量x并最终以exit(k max_iter ? 0 : 1)返回进程退出码——收敛则 0超限则 1方便脚本化校验。结果验证CPU 端残差检查收敛后程序在 CPU 端独立复算一遍稀疏矩阵向量乘rsum Σ val[j] * x[J[j]]并与右端项rhs[i]比较统计最大绝对误差errconjugateGradientCudaGraphs.cu最终打印Test Summary: Error amount %f。由于矩阵是三对角对称正定结构CG 在浮点精度内应快速收敛该误差即为对 GPU 求解正确性的端到端验证。涉及的 CUDA API 清单与资源管理README 列出的 CUDA Runtime API 全部可在源码中找到对应调用覆盖 Graph 全生命周期与流管理Graph 生命周期cudaStreamBeginCapture、cudaStreamEndCapture、cudaGraphInstantiate、cudaGraphLaunch、cudaGraphExecDestroy、cudaGraphDestroy流管理cudaStreamCreate、cudaStreamDestroy、cudaStreamSynchronize内存管理cudaMalloc、cudaFree、cudaMallocHost、cudaFreeHost、cudaMemcpyAsync含 H2D、D2D、D2H 三种方向、cudaMemsetAsync设备查询与内核配置cudaGetDeviceProperties、cudaOccupancyMaxPotentialBlockSize。程序退出前严格逆序释放资源先销毁 Graph 执行体与图对象、销毁流再cusparseDestroy/cublasDestroy销毁库句柄随后cusparseDestroySpMat/cusparseDestroyDnVec销毁描述符最后释放全部主机端页锁定内存与设备端显存conjugateGradientCudaGraphs.cu可作为 CUDA Graph 样例中资源管理范式的参考。小结conjugateGradientCudaGraphs 是一个麻雀虽小、五脏俱全的 CUDA Graph 教学样例它以 100 万阶三对角稀疏线性系统的 CG 求解为场景完整串联了 CUSPARSE 的 CSR 通用 APIcusparseSpMV、CUBLAS 的 BLAS-1 运算cublasSaxpy/cublasSdot/cublasSscal/cublasScopy以及 CUDA Graph 的捕获、实例化、回放与销毁全流程并提供了与非 Graph 路径的编译期开关WITH_GRAPH和 CPU 端结果校验是理解如何把多库混合的迭代算法 Graph 化以降低启动开销的理想起点。【免费下载链接】cuda-samplesSamples for CUDA Developers which demonstrates features in CUDA Toolkit项目地址: https://gitcode.com/GitHub_Trending/cu/cuda-samples创作声明：本文部分内容由AI辅助生成（AIGC），仅供参考