Que tipo de variáveis consomem registros no 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];
}
No exemplo acima, eu achox
, y
, offset
são salvos em registros enquanto
nvcc -Xptxas -v dá4 registers, 24+16 bytes smem
perfil mostra 4 registros
e o chefe deptx Arquivo:
.reg .u16 %rh<4>;
.reg .u32 %r<9>;
.reg .u64 %rd<10>;
.loc 15 21 0
$LDWbegin__Z3addPiPKiS1_:
.loc 15 26 0
Alguém pode esclarecer o uso de registros? No Fermi, o número máximo de registros é 63 para cada thread. No meu programa, quero testar o caso quando um kernel consome muitos registradores (portanto, as variáveis podem ter que ser armazenadas na memória local automaticamente e, assim, levar à diminuição do desempenho). Então neste ponto eu posso dividir um kernel em dois para que cada thread tenha registros suficientes. Suponha que os recursos do SM sejam suficientes para kernels simultâneos.
Não tenho certeza se estou certo.