为什么对推力::inclusive_scan 的调用比后续调用慢得多？答案

【问题标题】：Why the call to thrust::inclusive_scan is much slower than subsequent calls?为什么对推力::inclusive_scan 的调用比后续调用慢得多？
【发布时间】：2017-04-03 03:40:13
【问题描述】：

当我多次调用thrust::inclusive_scan时，为什么第一次比后面的调用慢很多？

这里是代码

float ttime;
for(int i=0;i<5;i++){
     cudaEvent_t start,stop; 
     cudaEventCreate(&start); 
     cudaEventCreate(&stop);
     cudaEventRecord(start,0);

     thrust::device_ptr<int > din(device_input);
     thrust::device_ptr<int > dout(device_output);
     thrust::inclusive_scan(din,din+N,dout);

     cudaEventRecord(stop,0); 
     cudaEventSynchronize(stop); 
     cudaEventElapsedTime(&ttime,start,stop);
     printf("cost %fms\n",ttime);
}

我在GTX1080上运行，结果是

cost 39.180702ms
cost 0.200704ms
cost 0.201728ms
cost 0.202752ms
cost 0.197632ms

谁能帮忙解释一下？

【问题讨论】：

标签： cuda thrust

【解决方案1】：

Thrust 是使用 CUDA 运行时 API 构建的，该 API 使用惰性 context initialisation。

没有记录确切的初始化顺序，并且有经验证据表明它随着时间的推移而改变。但是，上下文设置似乎是在ad hoc basis 上完成的。

第一次调用缓慢可能与程序中包含推力代码的模块的加载和初始化有关。您可以通过分析您的代码并查看配置文件执行时间与第一次调用的挂钟时间来验证这一点。

【讨论】：

【解决方案2】：

在 Talonmies 的有效答案中添加几句话：

在我的 this question 中，有一些关于 CUDA 初始化需要多少时间的鸡尾酒餐巾纸计算。
我还建议，为了将 libthrust 的加载+初始化时间与运行时 API 初始化开销分开，您执行以下三个阶段：
1. 执行一些不写任何输出的虚拟内核 - 两次（T_1，T2_）
2. 在（几乎）没有数据的情况下（但确实会启动内核）进行 libthrust 调用 - 两次（T_3、T_4）
3. 现在是您真正的通话时间了

粗略地说，(T_1 - T_2) 是 CUDA 负载和启动时间，(T_3 - T_4) 是推力负载和启动时间。

查看分析时间线很有帮助； CUDA 会将其大部分初始化“推送”到您的一个 API 调用中 - 但不一定是第一个。

【讨论】：