程序刚开始运行时,帧数显示为40fps,10秒左右后就变成5fps

我的程序刚开始运行时,帧数显示为40fps,仅过10秒左右,就变成5fps,请问可能的原因是什么?我交待一下我的程序运行过程:
首先我会创建一个数据集,然后用opengl将数据集渲染到两个纹理中,然后映射到CUDA进行使用。这个过程是动态变化的,也就是说,渲染到纹理的过程是持续不断的。在CUDA的核函数里,读取纹理并进行计算,最后生成另一个纹理,返回到opengl进行显示。
我基本可以肯定是kernel函数的问题,因为如果不运行核函数,而只进行其它操作的话,帧数不受影响,只有运行了核函数以后,帧数发生了明显减小

但如果说是kernel函数耗费时间的话,为什么一开始会有比较高的帧数呢?影响帧数的因素都有哪些呢?

请大家赐教

附,我的核函数源码如下:global void
d_render(uint *d_output, uint imageW, uint imageH,
float density, float brightness,
float transferOffset, float transferScale)
{
const int maxSteps = 500;
const float tstep = 0.01f;
const float opacityThreshold = 0.95f;

uint x = blockIdx.xblockDim.x + threadIdx.x;
uint y = blockIdx.y
blockDim.y + threadIdx.y;
if ((x >= imageW) || (y >= imageH)) return;

float u = (x / (float) imageW);
float v = (y / (float) imageH);

// calculate eye ray in world space
Ray eyeRay;
eyeRay.o = make_float3(mul(c_invViewMatrix, make_float4(0.0f, 0.0f, 0.0f, 1.0f)));
eyeRay.d = normalize(make_float3(u-0.5f, v-0.5f, -1.0f));
eyeRay.d = mul(c_invViewMatrix, eyeRay.d);

float4 sum = make_float4(0.0f);
float3 step = eyeRay.d*tstep;

float4 startPosition = tex2D(cudafronttex,u,v);//在opengl中生成的纹理,是持续生成的,即其内容是不断变化的,与这个有无关系?
float4 endPostion = tex2D(cudabacktex,u,v);//在opengl中生成的纹理
float3 pos;
pos.x = startPosition.x;
pos.y = startPosition.y;
pos.z = startPosition.z;

float depth; //下面这个求根,本来我有所怀疑,但用一个常量赋值实验后,帧数仍然发生减小
depth = sqrt((endPostion.x-pos.x)(endPostion.x-pos.x) + (endPostion.y-pos.y)(endPostion.y-pos.y) + (endPostion.z-pos.z)*(endPostion.z-pos.z));

float t = 0;
for(int i=0; i<maxSteps; i++) {
// read from 3D texture
float sample = tex3D(tex, pos.x, pos.y, pos.z);
// lookup in transfer function texture
float4 col = tex1D(transferTex, (sample-transferOffset)*transferScale);
col.w *= density;

// pre-multiply alpha

col.x *= col.w;
col.y *= col.w;
col.z = col.w;
// “over” operator for front-to-back blending
sum = sum + col
(1.0f - sum.w);

// exit early if opaque

if (sum.w > opacityThreshold) break;

t += tstep;
if (t > depth) break;
pos += step;
}
sum *= brightness;

// write output color生成新纹理并返回opengl
d_output[y*imageW + x] = rgbaFloatToInt(sum);
}

这个更大的可能是你的FPS算法的问题,有些计算FPS/计算每秒流量/计算需要在一定时间内累计工作量然后求平均工作量的算法都可能有类似问题。

真要确定是谁的问题,可以这样:

您跑下profiler, 看下运行400次以后贵kernel的时间有无变化,如果有,证明是kernel的问题;如果没有,可能是您的code的其他部分(例如上文说的计算FPS的算法)的问题。

再例如,一些频繁分配、释放内存+自动GC的语言、平台,也会遇到类似的越来越慢的现象。等等。

建议先看下您的kernel的时间变化,您觉得呢?

你好,关于FPS算法,我需要说明一下,这个代码是我用CUDA示例程序中的一个volume render程序改写过来的,所以FPS算法是照搬的原来的算法,我没有改动,其具体工作原理我也不是很清楚。
关于您给我的建议,我会尝试试一下,其实我还不会用profiler,我先自己搞一下,如果不会,再向您请教。
谢谢

在运行profiler测试的时候,好像遇到了out of memory错误 :frowning:

建议您使用nsight调试一下您的程序,看看具体是什么问题。

祝您调试顺利~

为什么我运行profiler好久也不出结果呢?总是显示正在运行以得到结果

LZ您好:

您的程序中是否有等待输入 的地方,如果有的话,会导致程序无限等待的。
请您参考一下如下链接:
http://cudazone.nvidia.cn/forum/forum.php?mod=viewthread&tid=6938

祝您好运~

你好,谢谢你的提醒。

我的程序是和opengl结合的,确实没有一个明确的出口,除非点击窗口的关闭按钮,否则会一直循环执行,这种情况下如何解决呢?

LZ您好:

请您设法修改您的代码,使程序能运行若干时间以后自动退出。

祝您好运~

你好,kernel的时间变化怎么看出来?而且我一直感觉是内存的问题,请问有没有这方面的思路?

你好,我一直感觉是内存的问题,请问有什么思路来看一下?

你好,我一直感觉是内存的问题,请问有没有思路来分析一下

LZ您好:

kernel的运行时间可以在profiler的时候统计出来,所以一直建议您运行profiler看看您kernel每次运行的时间有否变化。然后根据2#的建议隔离您问题可能出现的方面,再继续寻找。

在跑profiler的时候,需要反复运行您的程序多次,如果您的程序无法在运行确定时间/循环次数的时候结束,而是一直运行下去的话,profiler会一直跑着的,就死循环了。所以,建议您修改您的代码,以避免无限制地运行下去。

以上是对上文建议的总结,供您参考。

你好,谢谢你的回答

按照你的建议,我更改了一下程序,让程序生成一副画面就退出了,profiler共运行了37次,然后生成一个detail表,但那里只列出了我的核函数的开始时间(start time)和持续时间(duration time),我不知道每次的运行时间从哪里看,你能告诉我吗

LZ您好:

如果您使用的是CUDA 5自带的profiler,那么会直接看到一个timeline,可以很直观地看到各个环节的耗时和总的时间。这个是在建立session的时候直接会跑一次程序,建立这个timeline。之后可以选择更为细致的分析任务。

以及,如果是分析多次运行的情况下,kernel会否越来越慢,您可以修改您的代码,使您每次运行该程序,kernel被调用多次,然后跑profiler,您可以看看后期kernel会否明显变慢,再进一步寻找问题。

你好,请问我在程序里手动多次运行核函数,也可以吧,我用了一个for循环来调用核函数

LZ您好:

1:您在您的程序里面如何使用都是可以的,都可以跑profiler,也包括手工重复启动kernel。

2:上述1:的情况不保证能重现您原来程序的问题,因为他们是不同的程序。

我在程序里循环调用了40次,duration time没有明显变化,一直在186.3ms左右

LZ您好:

1:您如果是手工直接重复kernel的,那么您需要考证一下这样和您之前的用法相比,两者是否等价。之前的方法中是否会引入其他因素影响速度。

2:如果您是按照您原来的程序原样循环了40次,而kernel执行时间并无增加,那么您可以根据2#的建议,继续寻找问题,比如是FPS算法的方面。

3:以及严格地说,您这里kernel执行40次是否已经足够多,这个也需要您自己衡量一下,因为您之前的现象是10秒以后FPS才下降的,如果这段时间执行kernel的次数远多于40次,那么您或许需要考虑下加大您的测试规模。