system
1
问题一、
我要在下面这段代码中，每一次迭代对vector_a进行写操作，然后把vector_a中内容读出到cpu。vector_a是要重复利用的，要在下一次写入前读操作已经完成。我想实现memcpy和内核计算同时进行，节约时间，应该用什么方法实现这个需要。
for（int i=0;i<1000;i++）
{
kernel_1<<<>>>(*vector_a);//对vector_a 有写操作
cudaMemcpy(*host_b,*vector_a,d2h);//把vector_a的内容读出到cpu
kernel_2<<<>>>();//和vector_a无关
kernel_3<<<>>>();//和vector_a无关
}
我尝试用 stream来实现，
cudaStream_t stream;
cudaStreamCreate(&stream);
for（int i=0;i<1000;i++）
{
cudaStreamSynchronize(stream);
kernel_1<<<>>>(*vector_a);//对vector_a 有写操作
//cudaMemcpy(*host_b,*vector_a,d2h);//把vector_a的内容读出到cpu
cudaMemcpyAsync(host_a,k_grid_recsize,cudaMemcpyDeviceToHost,stream);//host_a 是用cudaMallocHost分配
// 的锁页内存
kernel_2<<<>>>();//和vector_a无关
kernel_3<<<>>>();//和vector_a无关
}
结果是memcpy和内核计算时间并不重叠，而且cudaMemcpyAsync所用时间是毫秒级别，太大。
问题二、我的一个cuda程序，一般都能运行处正确结果，但是有的时候（没有开其他cuda程序），程序启动后，在第一个cudaMalloc处就出现错误，说是out of memory（我分配的内存很小），重启电脑后，就又可以正常运行，这是什么原因呢
system
2
1：我觉得把kernel_1和memcpy放在一个流里面，kernel_2和kernel_3放在另外一个流里面，不需要用流同步即可。
2：尚不清楚为何会出现cudaMalloc的错误。
坐等横扫千军斑竹详细回答，我先打个前站。
system
3
楼主您好，您没有说明您的硬件是否支持复制和计算overlapping。 同时我看到您只创建了一个stream, 而且代码的最关键的部分：您的kernel是在哪个流上启动的，被您省略了。所以无法直接给出确切的答案。
根据您已有的信息，我作出如下假设：
（1）您的stream, 被kernel_1, cudaMemcpyAsync, kernel_2, kernel_3都使用了。
（2）您的kernel_1和cudaMemcpyAsync使用了您刚刚创立的stream, 而kernel_2和kernel_3没有使用任何显式的流。
如果您对您的流的使用情况如（1）猜测，那么自然是串行的，不能让您的计算和复制overlap。
如果您对您的流的使用情况如（2）猜测，那么您的后2个kernel在使用默认流，而使用默认流会导致一系列的自动的同步问题，从而可能导致您的在您的某个流上的复制操作和在默认流的计算无法同时进行。
对（2）猜测给出的建议：
楼主您可以尝试创建2个流，例如cudaStreamCreate(&你的第二个流),
然后将kernel_1和cudaMemcpyAsync放入您的第一个流；然后将您的kernel_2和kernel_3放入您的第二个流。这样应该是可以的。
额外的我建议楼主您提供更详细的信息。如果您认为还需要继续深入讨论。
关于楼主的帖子的第二部分，
您问道：为何有的时候会cudaMalloc失败，而此时却没有其他cuda程序使用显存。
从我个人的经验来说，一些普通的桌面程序会吃掉一部分显存，我曾经惊讶的发现某个版本的.p*f格式的阅读器会吃掉上百MB的显存。
所以对于第二部分我建议您检查下是否有其他桌面程序可以关闭。
需要说明的是，现在的显卡显存都很大。这个问题我只在老的128MB的我的一张老卡上见过。如果不是这个原因，欢迎楼主继续提供情况，让论坛和您一起分析。
system
4
LZ你好，
从你的代码来看，你只有一个流，那么无法再memcpy和kernel计算时overlap，所以需要至少两个流才可以，一个流负责与vector_a有关的写入和拷贝，另一个流来计算与vector_a无关的kernel计算。
第二种情况我建议可以在分配以前先对设备进行重置cudaDeviceReset(),如果还是会报错，可能就和楼上版主说的，你的显存被其他程序占用导致分配不到空间了！
system
5
我的kernel_1/kernel_2和kernel_3都是使用默认流，就cudaMemcpyAsync用的新建立的流。和你说的第二种情况类似。
我的显卡是GForce GT 240 ，128M显存。
新手发帖，情况没有描述清楚，多多包涵，非常感谢版主亲自回复
system
6
谢谢各位热心的朋友。我有个疑惑，我用cudaStreamCreate(&stream)建立了一个流，是不是还有一个默认的流？
system
7
从你的显卡来说，1.2的计算能力设备支持拷贝和计算的overlap,我觉得你可以创建两个流来代替默认流进行测试。
当然还有更重要的一点，Programming guide上指出了，实现拷贝和计算的overlap必须在host使用page-locked的分配内存，这点楼主也稍微注意一下！
system
8
回复楼主的6楼帖子：
是的，还有一个逻辑上的默认流存在。
额外的我建议楼主尝试我和tianyuan08的建议。
如果您依然有问题，欢迎您继续跟贴。
system
9
建议按照回复中一致的建议方法，建立两个stream，将kernel_1和copy放在一个流里面，kernel_2和kernel_3放在另外一个流里面。
你这里叙述的方法将不能保证每次kernel_1更改vector内容，并输出至host的逻辑正确性。
同时，您的显存如果只有128MB的话，是需要考虑下横扫千军斑竹说的情况，看看是否被其他应用占据了过多的显存。不过，如果确实是GT240 显卡的话，似乎一般应该有512MB显存才对的。
system
10
GT240, 曾经多么让我魂牵梦萦的名字。当年买不起，攒钱大半年，最后买了GTX460.
当年多少次徘徊在网吧门口，看着他们用大红条幅写着：机器都是最新的GT240显卡。。。
嘟嘟。
system
11
善哉，GTX460无论是CUDA特性还是游戏性能，都可以秒杀GT240的。
GT240当时，因为是新工艺，要比效能大致相当的9800GT省电很多，而且计算能力也是1.3的，被誉为1.3时代CUDA入门首选。更低端的9600GSO等，都是计算能力1.1的老核心了，而同为1.3计算能力的GTX260,280等都是又贵又热的电老虎。
system
12
:lol我现在用的是GTX480，GTX580，GTX680