Czym jest pamięć współdzielona i dlaczego warto jej używać?

Pamięć współdzielona w kontekście GPU i CUDA to obszar pamięci dostępny dla wszystkich wątków w ramach jednego bloku wątków. Znajduje się ona na samej GPU i jest znacznie szybsza od pamięci globalnej, ponieważ charakteryzuje się niskim opóźnieniem i wysoką przepustowością.

Użycie pamięci współdzielonej pozwala na efektywną wymianę danych między wątkami bloku bez konieczności sięgania do wolnej pamięci globalnej. Jest to szczególnie ważne dla algorytmów uczenia maszynowego i obliczeń równoległych, gdzie często konieczne jest współdzielenie wyników pośrednich.

Przykład: przy implementacji konwolucji lub operacji macierzowych można najpierw załadować dane do pamięci współdzielonej, wykonać obliczenia, a następnie zapisać wynik do pamięci globalnej, co znacznie przyspiesza wykonanie.

Przykład kodu CUDA:

__shared__ float tile[32][32];

int tx = threadIdx.x;
int ty = threadIdx.y;

// Wczytanie danych z pamięci globalnej do pamięci współdzielonej
tile[ty][tx] = input[y * width + x];
__syncthreads();

// Użycie danych z pamięci współdzielonej do obliczeń
float val = tile[ty][tx] * 2.0f;

// Zapisanie wyniku z powrotem do pamięci globalnej
output[y * width + x] = val;

W ten sposób pamięć współdzielona pomaga zoptymalizować wydajność poprzez zmniejszenie liczby wolnych odwołań do pamięci globalnej.