system
1
下面的一段测试代码， for循环中的kernel_1和kernel_2大部分时间的间隔是微秒级别，但是有的时候就变成毫秒级别，在这段时间里，gpu什么都没有做，而且是每隔一段时间，就出现一次这样的情况。我用 Nvidia visual profile 查看的时间，
图片见附件。从图片中，可以看出，
异常一：第一个kernel_1启动后大概隔了1.5ms才启动kernel_2，
异常二：kernel_1和kernel_2正常运行大概25次，就又有一段空白区
望各位朋友帮忙看看问题在哪里？是软件还是硬件（GT240）的问题。
int m_nt=100;
int num=100;
float host_a,host_b,dev_a,dev_b;
host_a=(float )malloc(numsizeof(float));
host_b=(float )malloc(numsizeof(float));
cudaMalloc((void **)&dev_a,numsizeof(float));
cudaMalloc((void **)&dev_b,numsizeof(float));
for(int i=0;i<num;i++)
{
host_a[i]=1;
host_b[i]=2;
}
cudaMemcpy(dev_a,host_a,numsizeof(float),cudaMemcpyHostToDevice);
cudaMemcpy(dev_b,host_b,numsizeof(float),cudaMemcpyHostToDevice);
for(int i=0;i<m_nt;i++) //for循环是主要的测试区域
{
kernel_1<<<1,num>>>(dev_b,dev_a);
kernel_2<<<1,num>>>(dev_b,dev_a);
kernel_1<<<1,num>>>(dev_b,dev_a);
}
cudaMemcpy(host_a,dev_a,numsizeof(float),cudaMemcpyDeviceToHost);
cudaMemcpy(host_b,dev_b,numsizeof(float),cudaMemcpyDeviceToHost);
[/i][/i]
system
2
楼主您好，您的现象应该相当正常。
对您的现象作出如下猜测：
（1）对于您的第一个情况，一个kernel, 不是上去就能立刻启动的，它需要一定的时间来热身。
举个例子：一个典型的情况是当你的exe里没有实际可以执行的GPU代码的时候，此时drvier需要时间将通用的虚拟机指令即时编译(JIT)为GPU的可执行的代码, 这个过程需要时间。
（2）您的CPU可能不足够快；以及，您的kernel足够小，很快就结束，而此时分配给您的进程的时间片无法持续满足您的进程对连续启动小kernel的需要。
举个例子1：您的CPU型号较老/性能较差。CPU时间无法持续分配给一个进程，使得存在无法喂饱GPU的情况。
再举个例子2：您使用了古董的单核CPU，此时，在默认流(例如你的代码)里积攒了足够多的命令后，显卡驱动迫切需要CPU时间来和GPU交互，此时您的进程将被OS的抢占式调度剥夺CPU时间，导致此时您的进程没有继续运行，您就观察到了中间的断层。
最后在举个2例子：您机器上运行有其他高优先级(包含因为曾经处于过阻塞态而导致优先级动态被临时调高)的进程，他们可能会抢占您的这个进程的CPU可用时间片。或者您的机器上有其他使用显卡资源的程序，例如一个以30FPS刷新窗口的播放器程序，会占用您的GPU。
对您建议的解决方案：
（1）对与第一个问题，给您2种建议的解决方案
(1-a)建议无视kernel的初次启动，然后测试以2，3，…次的数据为准。这样可以规避例如jit导致的热身时间。
(1-b)在编译的时候，尽量使用符合您的实际情况的虚拟架构和真实架构，例如如果您用的是GTX480, 可以考虑选择compute_20,sm_20的arch参数。因为如果存在可用的GPU二进制代码，将越过jit过程。
(2)对您的第二个问题，此现象一般常被CUDA用户称为运行在CPU上的代码无法及时的喂饱GPU.
可以考虑如下建议：
（1）更换更快的CPU, 或者更换更多核心的CPU, 或者不要运行过多无关软件，尽量在系统安静状态下进行测试。
（2）将您的kernel改为较大的大小，使得不需要很频繁的去喂您的显卡。
祝您测试愉快！
如果上述猜测/解决方案不正确或者无法满足您的要求，欢迎继续跟帖。
system
3
此外，建议注意排版，您的楼主位帖子中的代码又丢失了一部分。
system
4
这个结果图让我感到很奇怪，从profile上来看似乎两个内核在同时计算一样，楼主可否给个放大的时间序列呢？
system
5
根据LZ的信息，启动kernel的时候，每个kernel只有1个block，每个block100个线程，这个并不算多。希望LZ公开测试kernel的详细内容，看看给GPU运算的强度够不够大，看是否是kernel和线程规模过于轻量级造成的。
system
6
[attach]2746[/attach]
贴张大图，两个kernel不是并行的。诚如各位所言，我的kerel确实是太小了，
global void kernel_1(float * out,float* in)
{
int i=threadIdx.x + blockIdx.x * blockDim.x;
//for(int j=0;j<100;j++)
out[i]+=2*in[i]*in[i];
}
global void kernel_2(float * out,float* in)
{
int i=threadIdx.x + blockIdx.x * blockDim.x;
//for(int j=0;j<100;j++)
out[i]+=3*in[i]*in[i];
}
不加for(int j=0;j<100;j++)的对应第一次的图，加了对应现在的图。
加了for循环kernel的计算量大大增加后，for迭代就多次后，空白段还是存在。我的cpu是AMD Athlon X2 245 2.90Ghz 双核，这个cpu应该还可以，应该不是cpu问题吧。是不是我的显卡GT 240 不够给力的原因
system
7
选择compute_20,sm_20的arch参数 编译时的参数设置，这个没有接触过，能否给点资料或者指引
system
8
第一部分是给tianyuan08的。
第二部分，加上for 100, 依然没用，依然太小。CPU还行，和我的差不多。我是2.7G的双核，也是k10。
不是显卡不够给力的原因，相反，可能你的显卡(GT240)过猛。
还有，有本事你听听ICE的，别上去来一个<<<1,100>>>，说GPU无法全负荷工作，这样来恶心人！！
你的CPU还行！但是你的显卡比它更行！你一个host线程在哪里喂这种大小的kernel! 真心不行！我和你说过了，要么你加大你的kernel, 要么你换更更猛的CPU, 使得你如此小的kernel, 一个线程直接喂死！
试试看吧！欢迎骂我！
system
9
如果是nsight配置的CUDA项目，在VS里面，右键工程——属性——配置属性——CUDA Runtime API——GPU——GPU Architecture 可以改，在其中添加sm_20即可。
不过你的GT240是1.3的卡，上述配置是为fermi准备的（话说nsight配置的项目，这里直接是填好的）。
kernel多整几个block，每个block多点线程，kernel里面访存和复杂的指数运算，三角运算，除法，开方什么的多整点。
我用比你还差的CPU喂GTX470，循环1000次都没明显空隙出现。
system
10
真心谢谢各位。我现在编写的一个cuda程序是根据以前项目的c程序改编的，看来问题也是在kernel太小这里了。这个关键问题解决就可以交差了。
Thanks with all my heart.