我的kernel函数需要在执行的过程中让所有的thread同步，但使用“__syncthreads();”语句只能做到block内部的同步，请问如何做到block之间的同步呢？
目前想到的办法，就是把kernel函数在每次需要同步的地方断开，分成一个个小kernel，由主机端分别调用。
但这样做就必须先把已经加载到shared memory中的数据存回global memory，调用下一个小kernel时再重新加载回去，似乎做了很多重复劳动，请问还有什么更有效地办法吗？
另外弱弱的问一下，这个论坛的搜索功能在哪？
预计这个问题应该有人问过，想直接搜一下，可没找到搜索的选项……
block間的同步只能藉由kernel的啟動、結束來同步。
block間是無法同步的，因為每個block不是同時執行的，所以不能同步。
除非除非是只有很少的block才能同步，請google這篇論文
Inter-Block GPU Communication via Fast Barrier Synchronization
但能多少個block可以同步，這跟硬體的架構是有關係的
論壇文章編號6740對於這個問題有一些討論
LZ您好：
2# iHakka网友正解！一并感谢iHakka网友回馈论坛！
以及1#提到必须重新读入shared memory的问题，确实是这样的，必须重新读回，除非您能修改算法什么的。
论坛以前一直没有搜索功能，后来好像添加了，但是我也一时找不到地方了。
建议您使用google搜索代替，比如搜索 TDR，那么在google里面输入 TDR site:cudazone.nvidia.cn，这样临时应急好了。
同时也希望论坛支持团队能更好地完善论坛功能，造福网友。
祝您一切顺利~
正好需要这个问题的答案,感谢楼上几位朋友.
版主,我借用这个楼问一下:
楼上兄弟说: block間是無法同步的，因為每個block不是同時執行的，所以不能同步。
那么,有没有这种极端的情况: 一共有3个block, GPU先执行 block2 → block1 → block0.
还是说, GPU一定是按block0 → block1 → block2 执行, 只不过有可能不是同时执行.
多谢了,我最近在研究如何划分block,搜索时发现这个帖子.
楼主您好，这个无法保证。
如果您至少有3个SM, 每个SM上都有1个blocks,那么此时，显然无法也不应该保证任何执行顺序。
（因为原因很简单，如果保证了严格的0->1->2的block顺序执行，那么显然同一时刻只有1个SM在工作，另外2个必须空闲。这显然是不合理的）
您觉得呢？
嗯,有些感悟,多谢玫瑰…
追问一下:
是这样,如果我打算给GPU分配N个 blocks, 但是这个N很大很大…超过了GPU能一次性同时执行的blocks数量, 我想这种情况下, GPU肯定只能分批次执行这 N 个blocks(这样说对吧?). 假如GPU把这N 个blocks 分成了2个组, N1(block0 - blockm), N2(blockm+1 - blockN).
我问题是: GPU在执行时, 是随机执行(N1,N2)组呢? 还是按照block编号的顺序, 先执行N1, 再来N2?
多谢啊…
楼主您好，目前是这样的：
首先上去要平铺满所有的SM, 编号为N的block分别在编号为N的SM上，这样如果一圈下来，SM上还能继续上BLOCK, 则2N的在N号SM上，2N+1的在N+1号SM上，依次类推。
当执行中，有个SM上的某block结束了，则给此SM继续上block，但是此时编号是什么规律，我没有做过测试（猜测是和上面的一样？），NVIDIA也没有公开过资料，就不能说了。
以及，任何情况下，这个行为都是无保证的，这个测试是当年在某卡上做的，不能保证将来此行为不改变，因为要用靠Block在SM上的顺序来使用任何的部分blocks间同步的措施都是不安全的，因此强烈建议您不要这么用。
唯一安全的方式是重开个kernel，并且总是建议这样。
感谢您的来访。
EDIT:修正了几个错别字。
好的,我就是需要Blocks之间妥同工作.
您的讲解清晰明了,受益良多,玫瑰V5~~:victory::handshake
我想追问一下，在一个多次循环（2000次循环）里面调用核函数（也可以在核函数里面每个线程多次循环，但是块之间同步的问题不能解决），就会设计到多次的主机到设别的数据拷贝，这无疑将消耗大量的时间，请问版主有没有什么比较好的思路？？
yuhuilcm您好：
循环启动kernel并不一定要伴随着反复和host复制数据的，您可以把数据留在device端。
祝您好运~
就是，为何每次都要复制回来再回去？这不是无用功吗？
不要拘泥形式。直接下个kernel就地用上个kernel的结果缓冲区即可。
请您认真思索这点，不要被一些不好的例子坑害。
版主说的是啊！！！
如果涉及到计算数据的返回 作为下一次循环的输入的话 是不是情况又不一样了？
请您稍微详细叙述一下您的情况，最好有简要的伪码示意。
代码的形式有点像下面这种情况
01.for(int i=0;i<2000;i++)
02.{
03. C[i] = A[i]+B[best];
04. best = function(C[i]);
05.}
其中A,B,C都是赋予了初始值的很大的数组，best初始值为0，best的值每次都要通过函数function计算得到
LZ您好：
您没有给出这段代码是host端的，还是kernel中的，以及你没有叙述这段代码打算如何并行化。
我将之看做是C语言的原型代码示意。
以及，无论您是每个线程完成这样的一个循环迭代，还是一个block完成这样的一个迭代（function比较复杂需要一个block完成），还是一个kernel完成这样的一个迭代（host端循环发布kernel，kernel主要用于并行处理复杂的function），您的result都可以放在device端，无需回拷即可被下次使用。
大致如此，祝您好运~
在我没有叙述清楚的情况下 版主还是给出了我想要的答案，多谢版主，我大概知道我应该怎样做了
再追问版主一个问题 我的显卡是GTX650TI ，devicequery.exe的运行结果如下所示
每个块可用共享寄存器的最大数量是48K，每个块可用最大寄存器数量为64K。
那么我可以定义的块的个数是多少呢？
我的显卡有4个SM 每个SM有192个SP 是不是代表着我的GPU最多能一次允许4各块同时运行？？？