title: CUDA: Thread synchronization in the same block
body: <p>Im trying write program in CUDA but I have problem with synchronization in the same block between threads.</p>
  
  <p>Here is model situation:</p>
  
  <pre><code> 10 __global__ void gpu_test_sync()
   11 {
   12     __shared__ int t;
   13     int tid = threadIdx.x;
   14
   15     t = 0;
   16     __threadfence();
   17     __syncthreads();
   18
   19     // for(int i=0; i&lt;1000000 &amp;&amp; t&lt;tid; i++); // with fuse
   20     while(t&lt;tid);
   21
   22     t++;
   23     __threadfence();
   24 }
   25
   26 void f_cpu()
   27 {
   28     printf("TEST ... ");
   29     int blocks = 1;
   30     int threads = 2;
   31     gpu_test_sync&lt;&lt;&lt; blocks , threads &gt;&gt;&gt;();
   32     printf("OK\n");
   33 }
  </code></pre>
  
  <p>If threads = 1, everything is ok. If threads > 1, infinite cycling.</p>
  
  <p>Why? Function __threadfence(); should make visible value of t variable for other threads.</p>
  
  <p>How I can solve it?</p>