¿Qué tipo de variables consumen los registros en 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];
}

En el ejemplo anterior, supongox, y, offset se guardan en registros mientras

nvcc -Xptxas -v da4 registers, 24+16 bytes smem

perfilador muestra 4 registros

y el jefe deptx expediente:

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

$LDWbegin__Z3addPiPKiS1_:   
.loc    15  26  0  

¿Alguien puede aclarar el uso de los registros? En Fermi, el número máximo de registros es 63 para cada hilo. En mi programa, quiero probar el caso cuando un kernel consume demasiados registros (por lo que las variables deben almacenarse en la memoria local automáticamente y, por lo tanto, conlleva una disminución del rendimiento). Entonces, en este punto, puedo dividir un kernel en dos para que cada hilo tenga suficientes registros. Supongamos que los recursos SM son suficientes para los núcleos concurrentes.

No estoy seguro si estoy en lo correcto.