title: CUDA - Memory Limit - Vector Summation
body: <p>I'm trying to learn CUDA and the following code works OK for the values N&lt;= 16384, but fails for the greater values(Summation check at the end of the code fails, c values are always 0 for the index value of i>=16384).</p>
  
  <pre><code>#include&lt;iostream&gt;
  #include"cuda_runtime.h"
  #include"../cuda_be/book.h"
  
  #define N (16384)
  
  __global__ void add(int *a,int *b,int *c)
  {
      int tid = threadIdx.x + blockIdx.x * blockDim.x;
      if(tid&lt;N)
      {
          c[tid] = a[tid] + b[tid];
          tid += blockDim.x * gridDim.x;
      }
  }
  
  int main()
  {
      int a[N],b[N],c[N];
      int *dev_a,*dev_b,*dev_c;
  
      //allocate mem on gpu
      HANDLE_ERROR(cudaMalloc((void**)&amp;dev_a,N*sizeof(int)));
      HANDLE_ERROR(cudaMalloc((void**)&amp;dev_b,N*sizeof(int)));
      HANDLE_ERROR(cudaMalloc((void**)&amp;dev_c,N*sizeof(int)));
  
      for(int i=0;i&lt;N;i++)
      {
          a[i] = -i;
          b[i] = i*i;
      }
  
      HANDLE_ERROR(cudaMemcpy(dev_a,a,N*sizeof(int),cudaMemcpyHostToDevice));
      HANDLE_ERROR(cudaMemcpy(dev_b,b,N*sizeof(int),cudaMemcpyHostToDevice));
      system("PAUSE");
      add&lt;&lt;&lt;128,128&gt;&gt;&gt;(dev_a,dev_b,dev_c);
  
      //copy the array 'c' back from the gpu to the cpu
  
      HANDLE_ERROR( cudaMemcpy(c,dev_c,N*sizeof(int),cudaMemcpyDeviceToHost));
  
      system("PAUSE");
      bool success = true;
      for(int i=0;i&lt;N;i++)
      {
          if((a[i] + b[i]) != c[i])
          {
              printf("Error in %d: %d + %d != %d\n",i,a[i],b[i],c[i]);
              system("PAUSE");
              success = false;
          }
  
      }
  
      if(success) printf("We did it!\n");
  
      cudaFree(dev_a);
      cudaFree(dev_b);
      cudaFree(dev_c);
  
      return 0;
  }
  </code></pre>
  
  <p>I think it's a shared memory related problem, but I can't come up with a good explanation(Possible lack of knowledge). Could you provide me an explanation and a workaround to run for the values of N greater than 16384. Here is the specs for my GPU:</p>
  
  <pre><code>General Info for device 0
  Name: GeForce 9600M GT
  Compute capability: 1.1
  Clock rate: 1250000
  Device copy overlap : Enabled
  Kernel Execution timeout : Enabled
  Mem info for device 0
  Total global mem: 536870912
  Total const mem: 65536
  Max mem pitch: 2147483647
  Texture Alignment 256
  MP info about device 0
  Multiproccessor count: 4
  Shared mem per mp: 16384
  Registers per mp: 8192
  Threads in warp: 32
  Max threads per block: 512
  Max thread dimensions: (512,512,64)
  Max grid dimensions: (65535,65535,1)
  </code></pre>