Jakie zmienne zużywają rejestry w 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];
}
W powyższym przykładzie chybax
, y
, offset
są zapisywane w rejestrach
nvcc -Xptxas -v daje4 registers, 24+16 bytes smem
profiler pokazuje 4 rejestry
i głowaptx plik:
.reg .u16 %rh<4>;
.reg .u32 %r<9>;
.reg .u64 %rd<10>;
.loc 15 21 0
$LDWbegin__Z3addPiPKiS1_:
.loc 15 26 0
Czy każdy może wyjaśnić użycie rejestrów? W Fermi maksymalna liczba rejestrów wynosi 63 dla każdego wątku. W moim programie chcę przetestować przypadek, gdy jądro zużywa zbyt wiele rejestrów (więc zmienne mogą być automatycznie zapisywane w pamięci lokalnej, co prowadzi do zmniejszenia wydajności). W tym momencie mogę podzielić jedno jądro na dwa, aby każdy wątek miał wystarczającą liczbę rejestrów. Załóżmy, że zasoby SM są wystarczające dla współbieżnych jąder.
Nie jestem pewien, czy mam rację.