程序执行时间不稳定

system · 2013 年10 月 17 日 14:19

版主们好：

我的程序正常执行时间是5min左右，可是有时候会很慢，十几分钟不等。
这个时候显示cpu利用率特别低，0或者1.

程序主要框架是：

for（；；）
{
·········· ；
cudamemcpy（）；

for（；；）
{
········ //调用kernel函数
}

cudamemcpy（）；
············；
}

会调用kernel函数 9k 多次。
[table]
[tr][td] [attach]3468[/attach] [/td][/tr]
[/table][attach]3467[/attach]

用profiler跑程序，显示只有40%多的时间在执行kernel，其余时间如上图蓝色之间的间隔。应该是cpu的执行时间，但是cpu只执行了简单的for循环，不至于会用和执行kernel差不多的时间。不知道为什么？
还有就是程序执行五分钟，但是profiler汇总的结果图里面为什么只有9s钟？怎么才可以看到完整的程序执行时间图？
主要问题是：
1.一般程序执行五分钟，但是有时候会有十几分钟。此时cpu利用率特别低。怀疑是cpu的调度问题。
2.怎么可以看到在profiler里面程序的执行时间？如上图只显示9s钟。

system · 2013 年10 月 17 日 14:51

楼主您好：

请在您确定不再使用GPU后，加入cudaDeviceReset()语句，以便观察是否GPU任务一共就经历了9秒。

如果您添加后，发现GPU任务也经历了5分钟。那么自然这样就解决了。
但如果您发现的确GPU只有9秒，剩下的4分51秒不知道干嘛去了，您不妨使用VS的profiling功能，查看您具体是那些host代码片段消耗掉了时间。

关于您的host代码长期CPU使用率为0. 这个问题我建议您和CUDA分析分开，直接调试执行您的代码，然后再经过了，例如3分钟后，在CPU使用为0，而你又不知道你的host code代码在干嘛的时候，使用VS的break all菜单项，暂停所有线程的执行，然后查看下您的各个线程都阻塞在哪里。

关于您询问的为何蓝色之间有间隔，您的图发的不完整，能否将放大图再大一点范围复制一些，以便让我们看看是中间的空闲部分是多长，以及看看是否有其他的传输任务在进行中。

感谢来访，请按要求操作。

system · 2013 年10 月 18 日 14:42

谢谢您！
我在程序结束的部分加了个cudaDeviceReset 然后profiler，图如下：好奇怪！！！显示程序执行9s后就空白了，然后在560s的时候，有一个runtime api 。然后程序才结束。

如果不用profiler跑，程序执行大概5分钟，用event测时或手表都差不多这个时间。可是用profiler就得560s，差不多10分钟了。而且不知道为什么只显示了只有9s在执行操作。

图好像现在传不上来了。明天能传的时候传上来。
谢谢您！

system · 2013 年10 月 19 日 02:57

您好！

好像是服务器的问题，我的图传不上来了。
我在cpu利用率为0的时候，break all 显示主线程停在了这儿：

template
inline host cudaError_t cudaLaunch(
T entry
)
{
return cudaLaunch((const char)entry);
//这儿
}

关于profiler只能看9s钟。加了cudaDeviceReset后，可以看到560s。但是只有9s是有内容的，其余都是空白。在560s时，显示有一个 runtime api 。

蓝色之间的间隔，时间是370-470us左右。而且这个时间段没有memcpy和runtime api函数。就是说图中这一段是空的。

图发不上来，只能这样描述了。感谢！

system · 2013 年10 月 19 日 04:53

(1)如果您前9s符合您的总发布任务的此则，则您的图现在的确说明，您只有9s的GPU任务。

(2)us级别的空闲很正常，你稍微一不小心就容易让GPU出现空闲的。考虑到您图的比例，您大约16次kernel的运行和400us差不多，也就是一次kernel才20us多点，这只要你稍微发布命令不及时就容易出现空闲状态的，以及，根据你图看，更像是你的host thread被OS强占调度而导致留下几百us的空挡的。这个你也无法干涉，不过你可以：
使用2个或以上的host thread, 通过2个流发布任务，以规避稍微不留神和OS调度带来的空挡。

(3)为何您的主线程会长期停留在某函数的返回的右大括号}处，而不继续往下执行，如果您的描述属实，则在我几十年的调试生涯中从未见过此现象，无法为您解释第三点。孤陋寡闻了。

感谢周末来访，请尝试点2的建议。

system · 2013 年10 月 19 日 08:58

版主您好，感谢您在周末还帮忙解答我的问题。

现在我的程序最重要的问题是解决有时候程序好像卡死的问题（cpu利用率为0），至于您给的第二点建议，我准备在解决了这个问题后再做优化（我要去看看多线程编程）。

我 break all 后，显示是这样，我可能描述错了。。。

最后那个9s钟。我怀疑是我的profiler使用不当。因为我的程序耗时基本上都是在显卡的计算上，不可能只有9s的。 kernel函数g_Calc应该调用13518000次，而profiler跑出来的图只调用了196800次。程序在profiler里执行了560s ，那么196800/9 * 560 = 1224500 基本接近g_Calc的调用次数。而且在13518000次g_Calc之后，还有另外的6次kernel调用，大概十几秒，这个在profiler之中都没有显示出来。是因为显卡计算都是无数个重复没显示还是我哪里设置错了？

只有196800次g_Calc调用，应该有13518000次的。

只显示了9s，在大概560s处有一个 runtime api 。

未加cudaDeviceReset的图

这是之前的几张图，我用网络图片的方式传上来了。

最后弱弱问一下版主，break all 之后，我应该怎么看线程在哪执行，或者程序卡在了哪里？:lol

system · 2013 年10 月 19 日 08:59

周末愉快！:lol

system · 2013 年10 月 20 日 08:31

版主，我发现一些问题：
我的程序是这样的：

外层循环1500次；
内层循环30次；
每次内层循环调用300次kernel函数；

我给每次内层循环编号输出，发现在cpu为0的情况下，程序一直输出，并没有停止。只是输出速度比较慢。

给每次外层循环计时：

当我在截图完了之后，发现输出成这样：
一次外层循环执行时间减少！在鼠标没有任何操作后，时间又恢复。

然后我在程序执行时，不停的打开网页。如下图：

差不多每次200ms，1500*200=5min，正好差不多是我程序的执行时间。

结论：在执行程序时，cpu利用率为0的情况下（偶尔出现）。用不停的打开网页或者截图操作的方式刺激CPU，程序会恢复正常。盼望版主给予指导，谢谢！

system · 2013 年10 月 20 日 09:53

话说楼主你还是发点代码吧。

我看了这几楼的帖子，楼主先让我明确这几点：
（1）在你跑profiler前，无论是“慢”执行还是快速的执行，结果是否都是正确的？
（2）如果第一个问题的答案是“是”，请发送您的代码，而不是你的文字描述和一堆看不到的图。

谢谢。

system · 2013 年10 月 20 日 10:07

单从楼主用文字描述的流程看：
for() {传输, 多次kernel启动, 传输}
如果正确的执行了，那么是无问题的。

但是楼主既然遇到了profiler中的怪现象，以及，运行时间上的怪现象。

我觉得最大的可能要先考虑下是否执行正确，只有确定了执行正确这个基础前提，才能再考虑是否去做profiling, 否则一个错误的代码的profiling结果无意义的。

如我在前文说，您的文字描述看，无问题，
那么能否发送贵代码到论坛？
也许您的一些代码违背了您的文字描述本意，或者在字里行间有不经意的小问题，但只有阅读代码才能理解他们。希望能发送代码。

感谢您的周末来访。

system · 2013 年10 月 20 日 10:54

版主您好：
结果都是正确的。
代码特别特别长，kernel函数都有400行。可不可以私信或者打包邮件方式发送？
谢谢版主回复！

system · 2013 年10 月 20 日 12:43

楼主您好，kernel既然结果正确，可暂不发，

请主要发下环绕kernel启动周围的host代码部分.

谢谢。

system · 2013 年10 月 20 日 13:42

版主您好：

代码贴上了。

现在的主要问题是：

在程序执行时（出现CPU利用率为0的情况下），kernel会执行的很慢。如果期间采用打开网页等办法刺激一下，
就可以暂时恢复。不知道为什么。
我在笔记本上运行程序时没出现这样的问题。会不会是CPU能力不行。

for (j=0;j<m_GridnumXm_GridnumZ;j++)
{
v=m_Velocity[j];
m_Velocity[j]=1.1;
cudaMemcpy(m_vecVelocity,m_Velocity,sizeof(float)m_GridnumXm_GridnumZ,cudaMemcpyHostToDevice);
for(m_ShotX=0; m_ShotX<m_GridnumX; m_ShotX++)
{
WaveRecord(m_ShotX,m_vecRecord); //调用kernel函数
}
cudaMemcpy(m_Record,m_vecRecord,sizeof(float)m_tnumm_receiver,cudaMemcpyDeviceToHost);

for(i=0; i<m_receiverm_tnum; i++)
{
a1=m_Record[i];
a2=m_Arecord[i];
ADerivative[i][j]=10(a1-a2)/v;
}
m_Velocity[j]=v;
}

////////////////////////////////////////////////////////////////
////////////////////////////////////////////////////////////////
void WaveR::WaveRecord(int ShotX , float * m_vecR )
{
cudaMemset(m_vecUpLongX,0,m_PMLm_Widesizeof(float));
cudaMemset(m_vecUpLongZ,0,m_PMLm_Widesizeof(float));
cudaMemset(m_vecDownLongX,0,m_PMLm_Widesizeof(float));
cudaMemset(m_vecDownLongZ,0,m_PMLm_Widesizeof(float));
cudaMemset(m_vecUpLongX0,0,m_PMLm_Widesizeof(float));
cudaMemset(m_vecUpLongZ0,0,m_PMLm_Widesizeof(float));
cudaMemset(m_vecDownLongX0,0,m_PMLm_Widesizeof(float));
cudaMemset(m_vecDownLongZ0,0,m_PMLm_Widesizeof(float));
cudaMemset(m_UUpLongX,0,m_PMLm_Widesizeof(float));
cudaMemset(m_UUpLongZ,0,m_PMLm_Widesizeof(float));
cudaMemset(m_UDownLongX,0,m_PMLm_Widesizeof(float));
cudaMemset(m_UDownLongZ,0,m_PMLm_Widesizeof(float));
cudaMemset(m_vecLeftShortX,0,m_PMLm_GridnumZsizeof(float));
cudaMemset(m_vecLeftShortZ,0,m_PMLm_GridnumZsizeof(float));
cudaMemset(m_vecRightShortX,0,m_PMLm_GridnumZsizeof(float));
cudaMemset(m_vecRightShortZ,0,m_PMLm_GridnumZsizeof(float));
cudaMemset(m_vecLeftShortX0,0,m_PMLm_GridnumZsizeof(float));
cudaMemset(m_vecLeftShortZ0,0,m_PMLm_GridnumZsizeof(float));
cudaMemset(m_vecRightShortX0,0,m_PMLm_GridnumZsizeof(float));
cudaMemset(m_vecRightShortZ0,0,m_PMLm_GridnumZsizeof(float));
cudaMemset(m_ULeftShortX,0,m_PMLm_GridnumZsizeof(float));
cudaMemset(m_ULeftShortZ,0,m_PMLm_GridnumZsizeof(float));
cudaMemset(m_URightShortX,0,m_PMLm_GridnumZsizeof(float));
cudaMemset(m_URightShortZ,0,m_PMLm_GridnumZsizeof(float));
cudaMemset(m_vecPast,0,m_GridWidethm_GridHightsizeof(float));
cudaMemset(m_vecNow,0,m_GridWidethm_GridHightsizeof(float));
cudaMemset(m_vecNext,0,m_GridWidethm_GridHightsizeof(float));
cudaMemset(m_vecPast1,0,m_GridWidethm_GridHightsizeof(float));
cudaMemset(m_vecNow1,0,m_GridWidethm_GridHightsizeof(float));
cudaMemset(m_vecNext1,0,m_GridWidethm_GridHightsizeof(float));
if (ShotX==0)
{
cudaMemset(m_vecR,0,m_receiverm_tnumsizeof(float));
}
for (m_t=2;m_t<m_tnum+2;m_t++)
{
int m_ThreadX = 32;
int m_ThreadZ = 2;
if(m_t%2==0)
{
c_Calc(
m_vecRecord ,m_vecVelocity ,
m_alphax ,m_alphaz,
m_vecPast ,m_vecNow,
m_vecPast1 ,m_vecNow1,
m_vecUpLongX0 ,m_vecUpLongZ0,
m_vecDownLongX0 ,m_vecDownLongZ0,
m_vecUpLongX ,m_vecUpLongZ,
m_vecDownLongX ,m_vecDownLongZ,
m_UUpLongX ,m_UUpLongZ,
m_UDownLongX ,m_UDownLongZ,
m_vecLeftShortX0 ,m_vecLeftShortZ0,
m_vecRightShortX0 ,m_vecRightShortZ0,
m_vecLeftShortX ,m_vecLeftShortZ,
m_vecRightShortX ,m_vecRightShortZ,
m_ULeftShortX ,m_ULeftShortZ,
m_URightShortX ,m_URightShortZ,
m_GridWideth ,m_GridHight,
m_numX ,m_GridnumZ,
m_t ,ShotX,
m_ThreadX ,m_ThreadZ
);
}
else
{
c_Calc(
m_vecRecord ,m_vecVelocity ,
m_alphax ,m_alphaz,
m_vecPast1 ,m_vecNow1,
m_vecPast ,m_vecNow,
m_vecUpLongX ,m_vecUpLongZ,
m_vecDownLongX ,m_vecDownLongZ,
m_vecUpLongX0 ,m_vecUpLongZ0,
m_vecDownLongX0 ,m_vecDownLongZ0,
m_UUpLongX ,m_UUpLongZ,
m_UDownLongX ,m_UDownLongZ,
m_vecLeftShortX ,m_vecLeftShortZ,
m_vecRightShortX ,m_vecRightShortZ,
m_vecLeftShortX0 ,m_vecLeftShortZ0,
m_vecRightShortX0 ,m_vecRightShortZ0,
m_ULeftShortX ,m_ULeftShortZ,
m_URightShortX ,m_URightShortZ,
m_GridWideth ,m_GridHight,
m_numX ,m_GridnumZ,
m_t ,ShotX,
m_ThreadX ,m_ThreadZ

);
}
}
}

system · 2013 年10 月 21 日 12:54

不能沉不能沉

system · 2013 年10 月 22 日 02:20

请勿在本版以自顶的方式和/或其他方式灌水，否则将移送至水区。

system · 2013 年10 月 22 日 05:23

如果您真的出现某时“kernel运行的很慢”，那么我也没有办法了。

从您的代码看（如果您的WaveRecord只是您WaveRecord<<<>>>()的笔误的话），
那么贵代码无任何问题，同时已知所有调用的结果都正确执行了（您的前几楼），
那么一个简单的cudaMemcpy, kernel<<<>>>多次, cudaMemcpy再回来的过程是常规用法，无问题的，
我无法解释您的奇怪现象。

（话说您看到我让您更新下显卡驱动的要求了么？尝试了么？）

谢谢。