CUDA与TensorRT(3)之CUDA stream&Event&NVVP

最新推荐文章于 2023-10-21 16:34:14 发布

lonely-stone

最新推荐文章于 2023-10-21 16:34:14 发布

阅读量724

点赞数 1

分类专栏：深度学习文章标签： python linux 计算机视觉

本文链接：https://blog.csdn.net/qq_42692305/article/details/128746327

版权

7 篇文章 5 订阅

订阅专栏

定义：cudaStream_t stream
创建：cudaStreaCreate(&stream)
数据传输时：cudaMemcpyAsyn(dst, src, size, type, stream)
kernel在流中执行调用时：kernel_name<<<grid,block,shareMemSize,stream>>>(argument list)
同步流和查询流是否完成：
- cudaError_t cudaStreamSynchronize(cudaStream_t stream)
- cudaError_t cudaStreamQuery(cudaStream_t stream)
流的销毁：cudaError_t cudaStreamDestroy(cudaStream_t stream)

需要注意的：

在这里插入图片描述

这其中要求数据量和计算量足够大，不然达不到上图最下面这个效果，也没有流的优势了。

但是上图中为什么H2D和D2H为什么没有重叠呢？
因为CPU和GPU之间的数据传输是经过PCIe总线的，PCIe上的操作是顺序的。
带有双工PCIe总线的设备可以重叠两个数据传输，但他们必须在不同呢的流和不同方向上。

要求：GPU算力3.5以上，即Kepler架构及以上
API：cudaError_t cudaStreamCreateWithPriority(cudaStream_t *pStream, unsigned int flags, int priority)
cudaError_t cudaDeviceGetStreamPriorityRange(int *leastPriority)
特点：只对kernel有效，较低的整数值表示较高的流优先级，很少用。

多流为什么会有效，流越多越好么？

计算密集型：耗时在计算，一次访存，数十次甚至上百次计算
访存密集型：耗时在访存，一次访存，几次计算
GPU一般处理简单可并行计算，大部分kernel都是访存密集型
CUDA加速，kernel合并，将小任务合并成大任务（但不是大任务变成更大哈），更有效。
在这里插入图片描述

单线程内，默认流的执行是同步的（本部分讨论的默认流是设置了多个流的情况）
但是在编译的时候加上–defual-stream per-thread 后，默认流的执行也是异步的了。
当然在这两种情况中，显示流一直是异步的。

多线程情况下，编译的时候不加上述参数的话，就是多线程共享一个默认流。
加了上述参数进行编译的话，就是每个线程有一个默认流。

在stream中插入一个事件，类似于打一个标记位，用来记录stream是否执行到当前位置。Event有两个状态，已被执行和未被执行。

定义：cudaEvent_t event
创建：cudaError_t cudaEventCreate(cudaEvent_t* event);
插入流中：cudaError_t cudaEventRecord(cudaEvent_t event, cudaStream_t stream = 0);
销毁：cudaError_t cudaEventDestroy(cudaEvent_t event);
同步和查询
- cudaError_t cudaEventSynchronize(cudaEvent_t event);
- cudaError_t cudaEventQuery(cudaEvent_t event);
进阶同步函数cudaError_t cudaStreamWaitEvent(cudaStream_t stream, cudaEvent_t event);