尧图网站建设 
cuda-samples asyncAPI 示例精讲用 CUDA 事件实现 GPU 计时与 CPU/GPU 异步执行重叠【免费下载链接】cuda-samplesSamples for CUDA Developers which demonstrates features in CUDA Toolkit项目地址: https://gitcode.com/GitHub_Trending/cu/cuda-samplesasyncAPI 是 CUDA Samples 中0_Introduction入门序列的经典示例它的核心主题只有一个利用 CUDA 事件Events同时完成 GPU 端精确计时与 CPU/GPU 执行重叠overlap。在本示例中一批 CUDA 调用异步拷贝 kernel 异步回拷被串行提交到 stream而 CPU 侧通过查询事件状态得知 GPU 是否完成从而在 GPU 忙碌期间并行执行自己的计算任务。读完本文你将掌握cudaMemcpyAsync、cudaEventRecord、cudaEventQuery、cudaEventElapsedTime的组合用法并理解页锁定内存pinned memory为何是异步传输的前提。示例概览一条完整的异步流水线示例源码位于 cpp/0_Introduction/asyncAPI/asyncAPI.cu主流程如下在主机端用cudaMallocHost分配页锁定内存在设备端用cudaMalloc分配显存并用cudaMemset将显存初始化为255创建两个 CUDA 事件start/stop在默认 streamstream 0上依次提交cudaMemcpyAsyncH2D→increment_kernel→cudaMemcpyAsyncD2H并用事件标记这条流水线的起点与终点主机端进入一个while (cudaEventQuery(stop) cudaErrorNotReady)轮询循环在等待期间执行 CPU 自己的计算事件就绪后调用cudaEventElapsedTime读出 GPU 耗时同时用主机端定时器sdkCreateTimer测出 CPU 侧调用这些 CUDA 函数本身花费的时间两者对比即可直观看到异步带来的收益最后用correct_output校验结果并释放所有资源。示例数据规模为n 16 * 1024 * 1024约 1600 万个int共 64 MBkernel 以 512 线程/块、n/512个块启动把数组每个元素加上常数26。异步执行模型为什么 CPU 可以先走一步CUDA 的所有流式调用stream-based call都是异步的cudaMemcpyAsync、kernel launch 等一旦提交即立即把控制权还给 CPUGPU 在后台按 stream 顺序执行。这意味着CPU 无需等待 DMA 拷贝或 kernel 完成可以继续发指令、做计算多个操作在同一 stream 内保持 FIFO 顺序执行本示例全部提交到 stream 0不同 stream 之间则可以并行执行这也是重叠overlap能够发生的前提。本示例演示的正是最简单也最典型的重叠场景GPU 执行流水线期间CPU 在同一进程内持续做自己的计算代码中表现为counter循环。源码注释明确写道Since CUDA stream calls are asynchronous, the CPU can perform computations while GPU is executing (including DMA memcopies between the host and device)asyncAPI.cu。页锁定内存异步拷贝的必要条件cudaMemcpyAsync要真正异步执行主机端内存必须是页锁定内存pinned / page-locked memory这正是源码使用cudaMallocHost而非普通malloc的原因int *a 0; checkCudaErrors(cudaMallocHost((void **)a, nbytes)); // 页锁定主机内存 memset(a, 0, nbytes); int *d_a 0; checkCudaErrors(cudaMalloc((void **)d_a, nbytes)); // 设备显存 checkCudaErrors(cudaMemset(d_a, 255, nbytes)); // 初始化为 255见 asyncAPI.cu。页锁定内存不会被操作系统换出DMA 引擎可以安全地直接访问它因此 H2D/D2H 拷贝才能真正脱离 CPU 介入在后台进行如果使用可分页内存cudaMemcpyAsync会退化为同步行为内部先拷贝到临时页锁定缓冲。顺带一提cudaMemset(d_a, 255, nbytes)先把设备端数据全部置为255保证校验函数能探测到kernel 是否确实修改了每一个元素而不是碰巧通过。最终校验逻辑见correct_outputasyncAPI.cu任何一个元素不等于value26都会打印错误信息并返回失败。事件GPU 流水线的里程碑与计时工具CUDA 事件cudaEvent_t可以被插入 stream 的任意位置充当 GPU 执行进度的时间戳与同步点。本示例完整覆盖了事件生命周期的五个关键 APIAPI作用源码位置cudaEventCreate创建事件句柄asyncAPI.cucudaEventRecord(start, 0)在 stream 0 中记录起点事件asyncAPI.cucudaEventRecord(stop, 0)在 stream 0 中记录终点事件asyncAPI.cucudaEventQuery(stop)非阻塞查询事件是否已完成asyncAPI.cucudaEventElapsedTime计算两个事件之间的 GPU 耗时毫秒asyncAPI.cu事件查询轮询让 CPU 在等待中干活核心的异步等待逻辑非常简短// have CPU do some work while waiting for stage 1 to finish unsigned long int counter 0; while (cudaEventQuery(stop) cudaErrorNotReady) { counter; } checkCudaErrors(cudaEventElapsedTime(gpu_time, start, stop));见 asyncAPI.cu。cudaEventQuery是非阻塞的事件未完成时返回cudaErrorNotReady此时 CPU 进入循环体继续干自己的活完成后返回cudaSuccess退出循环。这与阻塞式的cudaDeviceSynchronize/cudaStreamSynchronize形成鲜明对比——后者会直接挂起 CPU 直到 GPU 全部完成无法实现执行重叠。循环结束后cudaEventElapsedTime返回start到stop之间 GPU 实际执行流水线两次异步拷贝 kernel所花费的毫秒数。程序随后输出三组关键数据asyncAPI.cuprintf(time spent executing by the GPU: %.2f\n, gpu_time); printf(time spent by CPU in CUDA calls: %.2f\n, sdkGetTimerValue(timer)); printf(CPU executed %lu iterations while waiting for GPU to finish\n, counter);第一行GPU 端真实执行耗时事件测出单位毫秒第二行CPU 端调用这批 CUDA 函数所占用的墙钟时间主机端定时器测出第三行CPU 在等待 GPU 期间完成的迭代次数直观证明CPU 没有闲着。主机端计时器另一个维度的对比为了对比 CPU 侧视角示例还使用了 SDK 辅助定时器。定时器的创建、启动、停止、取值分别对应 helper_timer.h 中的sdkCreateTimerhelper_timer.h、sdkStartTimer、sdkStopTimer、sdkGetTimerValue该头文件又由 helper_functions.h 统一引入。它在 Windows 上基于StopWatchWin实现、在 Linux 上基于StopWatchLinux实现见 helper_timer.h计时单位统一为毫秒。计时区间覆盖从cudaProfilerStart()到cudaProfilerStop()之间的全部 CUDA 调用提交过程asyncAPI.cucheckCudaErrors(cudaProfilerStart()); sdkStartTimer(timer); cudaEventRecord(start, 0); cudaMemcpyAsync(d_a, a, nbytes, cudaMemcpyHostToDevice, 0); increment_kernelblocks, threads, 0, 0(d_a, value); cudaMemcpyAsync(a, d_a, nbytes, cudaMemcpyDeviceToHost, 0); cudaEventRecord(stop, 0); sdkStopTimer(timer); checkCudaErrors(cudaProfilerStop());由于所有调用都是异步提交CPU 侧这段代码几乎是瞬间返回的因此CPU time会显著小于GPU time——这正是异步执行收益的直接体现。cudaProfilerStart/cudaProfilerStop用于在 nvprof / Nsight 等性能分析器下圈定采集区间日常运行不受影响。资源管理与错误检查规范示例在收尾阶段严格释放了全部资源顺序与分配顺序对应asyncAPI.cucheckCudaErrors(cudaEventDestroy(start)); checkCudaErrors(cudaEventDestroy(stop)); checkCudaErrors(cudaFreeHost(a)); checkCudaErrors(cudaFree(d_a)); exit(bFinalResults ? EXIT_SUCCESS : EXIT_FAILURE);值得注意的两个工程实践checkCudaErrors宏每个 CUDA Runtime API 调用都被包裹在错误检查宏中。该宏定义于 helper_cuda.h出错时会打印错误枚举、出错函数名、文件名与行号并退出进程是 cuda-samples 系列统一的错误处理风格也是生产代码中每个 CUDA 调用都必须检查返回值的样板。设备选择程序通过findCudaDevice(argc, argv)定义于 helper_cuda.h自动挑选计算能力最优的 CUDA 设备并支持命令行--deviceN显式指定设备随后用cudaGetDeviceProperties读出设备名称打印asyncAPI.cu。构建与运行示例提供标准 CMake 构建配置 cpp/0_Introduction/asyncAPI/CMakeLists.txt关键点如下通过find_package(CUDAToolkit REQUIRED)依赖本机已安装的 CUDA Toolkit默认编译目标架构CMAKE_CUDA_ARCHITECTURES覆盖 75 / 80 / 86 / 87 / 89 / 90 / 100 / 110 / 120 等主流 SM 版本启用-lineinfo调试构建时替换为-G以支持 CUDA 调试工具启用CUDA_SEPARABLE_COMPILATION编译标准为 C17 / CUDA 17并链接仓库根目录的Common辅助头文件目录include_directories(../../../Common)。典型构建与运行方式在仓库根目录执行mkdir build cd build cmake ../cpp/0_Introduction/asyncAPI make -j ./asyncAPI程序成功运行会打印设备名称、GPU 执行耗时、CPU 侧调用耗时以及 CPU 在等待期间执行的迭代次数并以退出码EXIT_SUCCESS结束。支持平台与环境依据 cpp/0_Introduction/asyncAPI/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、Windows支持 CPU 架构x86_64、armv7l前置条件安装与平台对应的 CUDA Toolkit。涉及的 CUDA Runtime API 速查README 中列出本示例用到的全部 CUDA Runtime API整理如下便于按图索骥查阅官方文档类别API内存管理cudaMalloc、cudaFree、cudaMallocHost、cudaFreeHost、cudaMemset、cudaMemcpyAsync事件cudaEventCreate、cudaEventRecord、cudaEventQuery、cudaEventElapsedTime、cudaEventDestroy同步cudaDeviceSynchronize设备查询cudaGetDeviceProperties性能分析cudaProfilerStart、cudaProfilerStop小结从示例到实践asyncAPI 用约 140 行代码讲清了异步编程的四个关键动作分配页锁定内存cudaMallocHost保证异步传输可行、用事件标记流水线里程碑cudaEventRecord、非阻塞查询进度cudaEventQuery让 CPU 在等待期间继续工作、用事件计算 GPU 耗时cudaEventElapsedTime。这套事件 异步流的组合是 CUDA 并发与性能优化的地基后续无论是多 stream 重叠、CUDA Graphs还是流水线化数据搬运都是在这一模型上的延伸。对初学者而言运行一遍本示例并对比输出的GPU time与CPU time是理解异步执行收益最直观的一课。【免费下载链接】cuda-samplesSamples for CUDA Developers which demonstrates features in CUDA Toolkit项目地址: https://gitcode.com/GitHub_Trending/cu/cuda-samples创作声明：本文部分内容由AI辅助生成（AIGC），仅供参考 cuda-samples asyncAPI 示例精讲：用 CUDA 事件实现 GPU 计时与 CPU/GPU 异步执行重叠
cuda-samples asyncAPI 示例精讲用 CUDA 事件实现 GPU 计时与 CPU/GPU 异步执行重叠【免费下载链接】cuda-samplesSamples for CUDA Developers which demonstrates features in CUDA Toolkit项目地址: https://gitcode.com/GitHub_Trending/cu/cuda-samplesasyncAPI 是 CUDA Samples 中0_Introduction入门序列的经典示例它的核心主题只有一个利用 CUDA 事件Events同时完成 GPU 端精确计时与 CPU/GPU 执行重叠overlap。在本示例中一批 CUDA 调用异步拷贝 kernel 异步回拷被串行提交到 stream而 CPU 侧通过查询事件状态得知 GPU 是否完成从而在 GPU 忙碌期间并行执行自己的计算任务。读完本文你将掌握cudaMemcpyAsync、cudaEventRecord、cudaEventQuery、cudaEventElapsedTime的组合用法并理解页锁定内存pinned memory为何是异步传输的前提。示例概览一条完整的异步流水线示例源码位于 cpp/0_Introduction/asyncAPI/asyncAPI.cu主流程如下在主机端用cudaMallocHost分配页锁定内存在设备端用cudaMalloc分配显存并用cudaMemset将显存初始化为255创建两个 CUDA 事件start/stop在默认 streamstream 0上依次提交cudaMemcpyAsyncH2D→increment_kernel→cudaMemcpyAsyncD2H并用事件标记这条流水线的起点与终点主机端进入一个while (cudaEventQuery(stop) cudaErrorNotReady)轮询循环在等待期间执行 CPU 自己的计算事件就绪后调用cudaEventElapsedTime读出 GPU 耗时同时用主机端定时器sdkCreateTimer测出 CPU 侧调用这些 CUDA 函数本身花费的时间两者对比即可直观看到异步带来的收益最后用correct_output校验结果并释放所有资源。示例数据规模为n 16 * 1024 * 1024约 1600 万个int共 64 MBkernel 以 512 线程/块、n/512个块启动把数组每个元素加上常数26。异步执行模型为什么 CPU 可以先走一步CUDA 的所有流式调用stream-based call都是异步的cudaMemcpyAsync、kernel launch 等一旦提交即立即把控制权还给 CPUGPU 在后台按 stream 顺序执行。这意味着CPU 无需等待 DMA 拷贝或 kernel 完成可以继续发指令、做计算多个操作在同一 stream 内保持 FIFO 顺序执行本示例全部提交到 stream 0不同 stream 之间则可以并行执行这也是重叠overlap能够发生的前提。本示例演示的正是最简单也最典型的重叠场景GPU 执行流水线期间CPU 在同一进程内持续做自己的计算代码中表现为counter循环。源码注释明确写道Since CUDA stream calls are asynchronous, the CPU can perform computations while GPU is executing (including DMA memcopies between the host and device)asyncAPI.cu。页锁定内存异步拷贝的必要条件cudaMemcpyAsync要真正异步执行主机端内存必须是页锁定内存pinned / page-locked memory这正是源码使用cudaMallocHost而非普通malloc的原因int *a 0; checkCudaErrors(cudaMallocHost((void **)a, nbytes)); // 页锁定主机内存 memset(a, 0, nbytes); int *d_a 0; checkCudaErrors(cudaMalloc((void **)d_a, nbytes)); // 设备显存 checkCudaErrors(cudaMemset(d_a, 255, nbytes)); // 初始化为 255见 asyncAPI.cu。页锁定内存不会被操作系统换出DMA 引擎可以安全地直接访问它因此 H2D/D2H 拷贝才能真正脱离 CPU 介入在后台进行如果使用可分页内存cudaMemcpyAsync会退化为同步行为内部先拷贝到临时页锁定缓冲。顺带一提cudaMemset(d_a, 255, nbytes)先把设备端数据全部置为255保证校验函数能探测到kernel 是否确实修改了每一个元素而不是碰巧通过。最终校验逻辑见correct_outputasyncAPI.cu任何一个元素不等于value26都会打印错误信息并返回失败。事件GPU 流水线的里程碑与计时工具CUDA 事件cudaEvent_t可以被插入 stream 的任意位置充当 GPU 执行进度的时间戳与同步点。本示例完整覆盖了事件生命周期的五个关键 APIAPI作用源码位置cudaEventCreate创建事件句柄asyncAPI.cucudaEventRecord(start, 0)在 stream 0 中记录起点事件asyncAPI.cucudaEventRecord(stop, 0)在 stream 0 中记录终点事件asyncAPI.cucudaEventQuery(stop)非阻塞查询事件是否已完成asyncAPI.cucudaEventElapsedTime计算两个事件之间的 GPU 耗时毫秒asyncAPI.cu事件查询轮询让 CPU 在等待中干活核心的异步等待逻辑非常简短// have CPU do some work while waiting for stage 1 to finish unsigned long int counter 0; while (cudaEventQuery(stop) cudaErrorNotReady) { counter; } checkCudaErrors(cudaEventElapsedTime(gpu_time, start, stop));见 asyncAPI.cu。cudaEventQuery是非阻塞的事件未完成时返回cudaErrorNotReady此时 CPU 进入循环体继续干自己的活完成后返回cudaSuccess退出循环。这与阻塞式的cudaDeviceSynchronize/cudaStreamSynchronize形成鲜明对比——后者会直接挂起 CPU 直到 GPU 全部完成无法实现执行重叠。循环结束后cudaEventElapsedTime返回start到stop之间 GPU 实际执行流水线两次异步拷贝 kernel所花费的毫秒数。程序随后输出三组关键数据asyncAPI.cuprintf(time spent executing by the GPU: %.2f\n, gpu_time); printf(time spent by CPU in CUDA calls: %.2f\n, sdkGetTimerValue(timer)); printf(CPU executed %lu iterations while waiting for GPU to finish\n, counter);第一行GPU 端真实执行耗时事件测出单位毫秒第二行CPU 端调用这批 CUDA 函数所占用的墙钟时间主机端定时器测出第三行CPU 在等待 GPU 期间完成的迭代次数直观证明CPU 没有闲着。主机端计时器另一个维度的对比为了对比 CPU 侧视角示例还使用了 SDK 辅助定时器。定时器的创建、启动、停止、取值分别对应 helper_timer.h 中的sdkCreateTimerhelper_timer.h、sdkStartTimer、sdkStopTimer、sdkGetTimerValue该头文件又由 helper_functions.h 统一引入。它在 Windows 上基于StopWatchWin实现、在 Linux 上基于StopWatchLinux实现见 helper_timer.h计时单位统一为毫秒。计时区间覆盖从cudaProfilerStart()到cudaProfilerStop()之间的全部 CUDA 调用提交过程asyncAPI.cucheckCudaErrors(cudaProfilerStart()); sdkStartTimer(timer); cudaEventRecord(start, 0); cudaMemcpyAsync(d_a, a, nbytes, cudaMemcpyHostToDevice, 0); increment_kernelblocks, threads, 0, 0(d_a, value); cudaMemcpyAsync(a, d_a, nbytes, cudaMemcpyDeviceToHost, 0); cudaEventRecord(stop, 0); sdkStopTimer(timer); checkCudaErrors(cudaProfilerStop());由于所有调用都是异步提交CPU 侧这段代码几乎是瞬间返回的因此CPU time会显著小于GPU time——这正是异步执行收益的直接体现。cudaProfilerStart/cudaProfilerStop用于在 nvprof / Nsight 等性能分析器下圈定采集区间日常运行不受影响。资源管理与错误检查规范示例在收尾阶段严格释放了全部资源顺序与分配顺序对应asyncAPI.cucheckCudaErrors(cudaEventDestroy(start)); checkCudaErrors(cudaEventDestroy(stop)); checkCudaErrors(cudaFreeHost(a)); checkCudaErrors(cudaFree(d_a)); exit(bFinalResults ? EXIT_SUCCESS : EXIT_FAILURE);值得注意的两个工程实践checkCudaErrors宏每个 CUDA Runtime API 调用都被包裹在错误检查宏中。该宏定义于 helper_cuda.h出错时会打印错误枚举、出错函数名、文件名与行号并退出进程是 cuda-samples 系列统一的错误处理风格也是生产代码中每个 CUDA 调用都必须检查返回值的样板。设备选择程序通过findCudaDevice(argc, argv)定义于 helper_cuda.h自动挑选计算能力最优的 CUDA 设备并支持命令行--deviceN显式指定设备随后用cudaGetDeviceProperties读出设备名称打印asyncAPI.cu。构建与运行示例提供标准 CMake 构建配置 cpp/0_Introduction/asyncAPI/CMakeLists.txt关键点如下通过find_package(CUDAToolkit REQUIRED)依赖本机已安装的 CUDA Toolkit默认编译目标架构CMAKE_CUDA_ARCHITECTURES覆盖 75 / 80 / 86 / 87 / 89 / 90 / 100 / 110 / 120 等主流 SM 版本启用-lineinfo调试构建时替换为-G以支持 CUDA 调试工具启用CUDA_SEPARABLE_COMPILATION编译标准为 C17 / CUDA 17并链接仓库根目录的Common辅助头文件目录include_directories(../../../Common)。典型构建与运行方式在仓库根目录执行mkdir build cd build cmake ../cpp/0_Introduction/asyncAPI make -j ./asyncAPI程序成功运行会打印设备名称、GPU 执行耗时、CPU 侧调用耗时以及 CPU 在等待期间执行的迭代次数并以退出码EXIT_SUCCESS结束。支持平台与环境依据 cpp/0_Introduction/asyncAPI/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、Windows支持 CPU 架构x86_64、armv7l前置条件安装与平台对应的 CUDA Toolkit。涉及的 CUDA Runtime API 速查README 中列出本示例用到的全部 CUDA Runtime API整理如下便于按图索骥查阅官方文档类别API内存管理cudaMalloc、cudaFree、cudaMallocHost、cudaFreeHost、cudaMemset、cudaMemcpyAsync事件cudaEventCreate、cudaEventRecord、cudaEventQuery、cudaEventElapsedTime、cudaEventDestroy同步cudaDeviceSynchronize设备查询cudaGetDeviceProperties性能分析cudaProfilerStart、cudaProfilerStop小结从示例到实践asyncAPI 用约 140 行代码讲清了异步编程的四个关键动作分配页锁定内存cudaMallocHost保证异步传输可行、用事件标记流水线里程碑cudaEventRecord、非阻塞查询进度cudaEventQuery让 CPU 在等待期间继续工作、用事件计算 GPU 耗时cudaEventElapsedTime。这套事件 异步流的组合是 CUDA 并发与性能优化的地基后续无论是多 stream 重叠、CUDA Graphs还是流水线化数据搬运都是在这一模型上的延伸。对初学者而言运行一遍本示例并对比输出的GPU time与CPU time是理解异步执行收益最直观的一课。【免费下载链接】cuda-samplesSamples for CUDA Developers which demonstrates features in CUDA Toolkit项目地址: https://gitcode.com/GitHub_Trending/cu/cuda-samples创作声明：本文部分内容由AI辅助生成（AIGC），仅供参考