CUDA Academy · Lekcja

Rejestry i pamięć lokalna

Poznać najszybsze miejsce przechowywania danych wątku i jego przepełnienia.

Lekcja 1 z 413 kroki

Rejestry i pamięć lokalna to bezpłatna lekcja CUDA Academy na CoddyKit. To lekcja 1 z 4. Możesz przeczytać całą lekcję poniżej za darmo — a potem ćwiczyć ją interaktywnie w przeglądarce z wbudowanym edytorem kodu i tutorem AI dostępnym 24/7. To część ścieżki edukacyjnej CUDA Academy, a Twój postęp synchronizuje się między webem a aplikacją CoddyKit. Kurs CUDA Academy zawiera 4 lekcji w sumie.

Najszybsza pamięć, jaką masz

Każdy wątek otrzymuje własne, prywatne rejestry — najszybszą pamięć układu. Znajdują się tuż obok jednostek obliczeniowych, więc odczyt z nich prawie nic nie kosztuje.

Zwykłe zmienne stają się rejestrami

Gdy w jądrze zapisujesz zwykłą zmienną lokalną, kompilator zazwyczaj przechowuje ją w rejestrze. Nie jest do tego potrzebna żadna specjalna składnia. 🙂

__global__ void k() {
    int x = 5;   // x lives in a register
    float y = x * 2.0f;
}

Prywatne dla jednego wątku

Wartość w rejestrze należy dokładnie do jednego wątku. Wątek 7 nie może odczytać rejestru wątku 8, dlatego praca poszczególnych wątków pozostaje wyraźnie rozdzielona.

Rejestry tworzą wspólną pulę

Każdy multiprocesor strumieniowy ma stały plik rejestrów. Wszystkie aktywne wątki dzielą tę pulę, więc mniejsze zużycie rejestrów na wątek pozwala uruchomić jednocześnie więcej wątków.

Gdy zabraknie miejsca

Jeśli jądro potrzebuje więcej rejestrów, niż jest dostępnych, dodatkowe wartości są przenoszone gdzie indziej. To zdarzenie nazywa się przepełnieniem rejestrów.

Dokąd trafiają przepełnienia: pamięć lokalna

Wartości przeniesione w wyniku przepełnienia trafiają do pamięci lokalnej, która — mimo swojej nazwy — nie znajduje się w układzie. W rzeczywistości mieści się w wolnej, zewnętrznej pamięci DRAM.

Lokalna nadal oznacza prywatną dla wątku

Pamięć lokalna jest prywatna dla każdego wątku, podobnie jak rejestry. Różnica dotyczy wyłącznie szybkości, ponieważ dostęp do pamięci lokalnej trwa znacznie dłużej.

Duże tablice lokalne powodują przepełnienia

Zadeklarowanie dużej tablicy przypisanej do wątku często wymusza umieszczenie jej w pamięci lokalnej, ponieważ zwyczajnie nie ma wystarczającej liczby rejestrów, aby przechować każdy element.

__global__ void k() {
    float buf[64]; // likely lives in local memory
}

Przepełnienia obniżają wydajność

Każde przepełnienie zamienia bezpłatny dostęp do rejestru na powolny odczyt z DRAM. Ograniczenie zapotrzebowania na rejestry to jeden z najprostszych sposobów na optymalizację.

Sprawdzanie liczby rejestrów

Można polecić kompilatorowi raportowanie liczby rejestrów na wątek. Przekaż do nvcc opcję --ptxas-options=-v, a narzędzie wypisze zużycie dla każdego jądra.

nvcc --ptxas-options=-v kernel.cu

Celowe ograniczanie liczby rejestrów

Kwalifikator __launch_bounds__ podpowiada kompilatorowi, aby ograniczył liczbę rejestrów, poświęcając nieco szybkości pojedynczego wątku na rzecz jednoczesnego uruchomienia większej liczby wątków.

__global__ void __launch_bounds__(256)
myKernel() { /* ... */ }

Szybkie sprawdzenie

Zadeklarowano dużą tablicę przypisaną do wątku, a wydajność spadła. Co najprawdopodobniej się stało?

Odpowiedz w interaktywnej lekcji

Podsumowanie: szybko i prywatnie

Dowiedziałeś się, że rejestry są najszybszą pamięcią przypisaną do wątku, ich pula jest ograniczona, a nadmiar trafia do wolnej pamięci lokalnej. Ograniczaj zapotrzebowanie na rejestry. 🎯

Bezpłatny start

Ucz się C++ dzięki korepetycjom AI — za darmo

Pisz i uruchamiaj kod w przeglądarce, otrzymuj natychmiastową pomoc od korepetytora AI dostępnego 24/7 i kontynuuj naukę w sieci lub w aplikacji.

Kursy: 30
Lekcje: 120

Zacznij naukę bezpłatnie Wyświetl pełny kurs

Często zadawane pytania

Czy lekcja „Rejestry i pamięć lokalna” jest bezpłatna?

Tak — pełny tekst „Rejestry i pamięć lokalna” jest dostępny za darmo tutaj w sieci. Aby ćwiczyć ją interaktywnie (wbudowany edytor kodu i tutor AI dostępny 24/7) i odblokować resztę kursu CUDA Academy, przejdź na CoddyKit PRO. Kurs CUDA Academy zawiera 4 lekcji w sumie.

Co nauczysz się w „Rejestry i pamięć lokalna”?

Poznać najszybsze miejsce przechowywania danych wątku i jego przepełnienia. Ćwiczysz CUDA Academy z praktycznym kodem, który uruchamiasz bezpośrednio w przeglądarce, a tutor AI dostępny 24/7 odpowiada na Twoje pytania podczas pracy nad lekcją.

Czy potrzebuję doświadczenia, aby zacząć CUDA Academy?

Nie wymagamy żadnego doświadczenia. CUDA Academy w CoddyKit jest strukturyzowany dla początkujących i zaawansowanych użytkowników, więc możesz zacząć tutaj lub od początku i uczyć się w swoim tempie. To lekcja 1 z 4.

Ile czasu zajmuje lekcja „Rejestry i pamięć lokalna”?

Większość lekcji CoddyKit trwa około 5–10 minut. Każda lekcja to mały, interaktywny krok, dzięki czemu robisz systematyczne postępy i zawsze wracasz dokładnie do tego samego miejsca — na webie i w aplikacji.

Czy mogę pisać i uruchamiać kod w tej lekcji CUDA Academy?

Tak. Każda lekcja CUDA Academy zawiera wbudowany edytor kodu, więc piszesz i uruchamiasz prawdziwy kod bezpośrednio w przeglądarce i od razu otrzymujesz sprzężenie zwrotne od AI — bez konfiguracji na komputerze.

Wszystkie lekcje w tym kursie

← Powrót do CUDA Academy