Какие переменные потребляют регистры в CUDA?

__global__ void add( int *c, const int* a, const int* b )
{
    int x = blockIdx.x;
    int y = blockIdx.y;
    int offset = x + y * gridDim.x;
    c[offset] = a[offset] + b[offset];
}

В приведенном выше примере, я думаю, x, y, offsetсохраняются в регистрах, а

• nvcc -Xptxas -v дает4 registers, 24+16 bytes smem

• профайлер показывает 4 регистра

• и заголовок ptx файла:

  .reg.u16 %rh<4>;
  .reg.u32 %r<9>;    
  .reg.u64 %rd<10>;  
  .loc    15  21  0   
  
  $LDWbegin__Z3addPiPKiS1_:   
  .loc    15  26  0  
  

Может ли кто-нибудь прояснить использование регистров? В Fermi максимальное количество регистров равно 63 для каждого потока. В своей программе я хочу протестировать случай, когда ядро ​​потребляет слишком много регистров (, поэтому переменные могут автоматически сохраняться в локальной памяти, что приводит к снижению производительности ). Затем в этот момент я могу разделить одно ядро ​​на два, чтобы у каждого потока было достаточно регистров. Предположим, что ресурсов SM достаточно для параллельных ядер.

Я не уверен, прав ли я.