title: Shift operator in CUDA
body: <p>Given this CUDA code, I am trying to perform bit shifting operations and the return on these values is zero. This should not be happening. Does anyone know how to fix this issue? Am I missing a CUDA header include?</p>
  
  <p><strong>Code</strong></p>
  
  <pre><code>      __device__ unsigned int FI( unsigned int in_data, unsigned int subkey, 
       unsigned int *KLi1, unsigned int *KLi2, unsigned int *KOi1, unsigned int *KOi2,
       unsigned int *KOi3, unsigned int *KIi1, unsigned int *KIi2, unsigned int *KIi3) {
  
        unsigned int nine, seven;
  
        unsigned int S7[128]  = {};
  
        unsigned int S9[512] = {};
  
        nine = (in_data&gt;&gt;7);
        seven = (in_data&amp;0x7F);
  
        /* Now run the various operations */
        nine = (unsigned int)(S9[nine] ^ seven);
        seven = (unsigned int)(S7[seven] ^ (nine &amp; 0x7F));
        seven ^= (subkey&gt;&gt;9);
        nine ^= (subkey&amp;0x1FF);
        nine = (unsigned int)(S9[nine] ^ seven);
        seven = (unsigned int)(S7[seven] ^ (nine &amp; 0x7F));
        in_data = (unsigned int)((seven&lt;&lt;9) + nine);
        return( in_data );
        }
  </code></pre>
  
  <hr>
  
  <p><strong>Breakpoint Analysis</strong></p>
  
  <p>Here is an example of a code snippet that shifts an unsigned int 7 places to the right. When I cuda-gdb my exec and breakpoint at the instruction, I observe that the value after shifting remains zero when it shouldn't. When I normally execute the same operation in cuda-gdb command prompt, I get a non-zero value. Any suggestions or hints?</p>
  
  <p>The variables nine and seven should be non-empty based on the value of <code>in_data</code>.</p>
  
  <pre><code>    nine = (in_data&gt;&gt;7);
      seven = (in_data&amp;0x7F);
  
      [Switching focus to CUDA kernel 0, grid 1, block (0,0,0), thread (1,0,0), device 0, sm 0, warp 0, lane 1]
      Breakpoint 1, FI (KLi1=0x3fffae0, KLi2=0x3fffb00, KOi1=0x3fffb20, KOi2=0x3fffb40, KOi3=0x3fffb60, 
      KIi1=0x3fffb80, KIi2=0x3fffba0, KIi3=0x3fffbc0, in_data=461, subkey=0) at kasumiOp.cu:61
       61   nine = (in_data&gt;&gt;7);
       (cuda-gdb) p in_data
       $1 = 461
       (cuda-gdb) step
       62   seven = (in_data&amp;0x7F);
       (cuda-gdb) p nine
       $2 = 0
       (cuda-gdb) step
       65   nine = (unsigned int)(S9[nine] ^ seven);
       (cuda-gdb) p seven
       $3 = 0
       (cuda-gdb) p 461 &gt;&gt; 7
       $4 = 3
       (cuda-gdb) cuda thread
       thread (1,0,0)
       (cuda-gdb) p 561 &amp; 0x7f
       $5 = 49
       (cuda-gdb) p 461 &amp; 0x7f
       $6 = 77
  </code></pre>
  
  <p>So, <code>in_data</code> is a value. I will try a trivial example and see if I can reproduce the same. </p>