CUDA Academy · Lekcja

Warunki wyścigu na GPU

Zrozumieć, dlaczego równoczesne zapisy uszkadzają dane.

Lekcja 1 z 413 kroki

Warunki wyścigu na GPU 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.

Wiele rąk, jeden licznik

Na GPU tysiące wątków może jednocześnie korzystać z tej samej pamięci. Gdy kilka z nich wspólnie zapisuje w jedno miejsce, powstaje wyścig. 🏁

Czym naprawdę jest wyścig

Wyścig występuje, gdy końcowy wynik zależy od nieprzewidywalnej kolejności wykonywania wątków. Ten sam kod może dać inny wynik przy każdym uruchomieniu.

Odczyt-modyfikacja-zapis

Zwiększanie licznika to w rzeczywistości trzy kroki: odczyt wartości, dodanie jedynki i zapisanie wyniku. Właśnie w tej sekwencji odczyt-modyfikacja-zapis pojawia się problem.

counter = counter + 1;

Jak przeplatają się kroki

Dwa wątki odczytują 5, oba obliczają 6 i oba zapisują 6. Oczekiwano wartości 7, ale otrzymano 6. Jedno zwiększenie po cichu zniknęło.

Utracona aktualizacja

To zniknięte zwiększenie nazywa się utraconą aktualizacją. Gdy rywalizują tysiące wątków, jednocześnie może zniknąć dziesiątki lub setki aktualizacji.

Kernel, który wygląda poprawnie

Ten kernel wygląda poprawnie, ale każdy wątek konkuruje o ten sam adres. Końcowa wartość będzie nieprawidłowa i będzie się różnić między uruchomieniami.

__global__ void count(int* total) {
    *total = *total + 1;
}

Dlaczego wynik jest niedeterministyczny

Sprzęt nigdy nie gwarantuje kolejności wykonywania wątków. Dlatego kernel z wyścigiem jest niedeterministyczny: może nawet działać poprawnie dla małych danych, a zawodzić dla dużych.

Same odczyty są bezpieczne

Wiele wątków odczytujących tę samą wartość jest całkowicie bezpieczne. Problem zaczyna się dopiero wtedy, gdy co najmniej jeden wątek zapisuje, podczas gdy inne odczytują lub zapisują.

Rozłączne zapisy też są bezpieczne

Jeśli każdy wątek zapisuje do własnego, unikatowego miejsca, takiego jak out[i], nie ma konfliktu. Do wyścigu potrzebne są wątki kierujące zapis do tego samego miejsca.

out[i] = a[i] + b[i];

Zapowiedź rozwiązania

Rozwiązaniem jest wykonanie operacji odczyt-modyfikacja-zapis jako jednego niepodzielnego kroku. Jest to operacja atomowa, którą omówimy w następnej lekcji. ⚛️

Wykrywanie wyścigów podczas przeglądu kodu

Podczas przeglądania kernela proszę zapytać: czy dwa wątki zapisują pod ten sam adres bez ochrony? Jeśli tak, niemal na pewno występuje wyścig danych.

Szybkie sprawdzenie

Upewnijmy się, że idea wyścigu jest już jasna.

Podsumowanie: wyścigi na GPU

Dowiedziałeś się, że niezabezpieczona operacja odczyt-modyfikacja-zapis na współdzielonych danych powoduje wyścigi i utracone aktualizacje. Operacje atomowe, które omówimy za chwilę, czynią te kroki niepodzielnymi. ✅

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

Często zadawane pytania

Czy lekcja „Warunki wyścigu na GPU” jest bezpłatna?

Tak — pełny tekst „Warunki wyścigu na GPU” 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 „Warunki wyścigu na GPU”?

Zrozumieć, dlaczego równoczesne zapisy uszkadzają dane. Ć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 „Warunki wyścigu na GPU”?

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

  1. Warunki wyścigu na GPU
  2. atomicAdd i funkcje pokrewne
  3. Budowanie histogramu
  4. Własne operacje atomowe z atomicCAS
← Powrót do CUDA Academy