CUDA Academy · Lektion

Cooperative Groups

Eine moderne API für flexible Synchronisierungsbereiche.

Lektion 4 von 413 Schritte

Cooperative Groups ist eine kostenlose CUDA Academy-Lektion auf CoddyKit. Dies ist Lektion 4 von 4. Du kannst die komplette Lektion unten kostenlos lesen – dann übst du sie direkt im Browser mit einem integrierten Code-Editor und einem KI-Tutor rund um die Uhr. Sie ist Teil des CUDA Academy-Lernpfads, und dein Fortschritt wird über Web und CoddyKit-App synchronisiert. Der CUDA Academy-Kurs umfasst insgesamt 4 Lektionen.

Eine übersichtlichere Synchronisierungs-API

Rohe Masken und Intrinsics funktionieren, sind aber umständlich. Cooperative Groups fassen Threads in Objekten zusammen, die Sie benennen, deren Größe Sie bestimmen und die Sie explizit synchronisieren können.

#include <cooperative_groups.h>
namespace cg = cooperative_groups;

Den gesamten Block erfassen

Holen Sie sich zunächst eine Referenz auf Ihren Thread-Block. this_thread_block gibt eine Gruppe zurück, die Sie wie syncthreads synchronisieren können, allerdings als Objekt.

cg::thread_block block = cg::this_thread_block();

Über die Gruppe synchronisieren

Der Aufruf von sync für den Block ist die moderne Barriere. Er bewirkt genau dasselbe wie syncthreads, liest sich aber klar als Methode der gemeinten Gruppe.

block.sync();

Eine Warp-Kachel abteilen

Sie können einen Block in Kacheln fester Größe aufteilen. Eine Kachel mit 32 Lanes ergibt eine Gruppe in Warp-Größe mit übersichtlichen Methoden statt roher Shuffle-Masken.

auto warp = cg::tiled_partition<32>(block);

Methoden ersetzen Masken

Eine gekachelte Gruppe stellt shfl_down und verwandte Funktionen ohne Maskenargument bereit. Die Gruppe kennt ihre Mitglieder bereits, daher bleibt die API kurz und sicher.

val += warp.shfl_down(val, offset);

Die eigene Position kennen

Jede Gruppe stellt ihre Mitglieder und Ihre Position bereit. thread_rank liefert Ihren Index innerhalb der Gruppe, und size gibt an, wie viele Threads sie enthält.

int rank = warp.thread_rank();

Auch kleinere Kacheln

Kacheln müssen nicht 32 Lanes breit sein. Eine tiled_partition mit 8 oder 16 Lanes erzeugt Sub-Warp-Gruppen – praktisch, wenn Ihre Daten natürlicherweise in kleinen Mengen gebündelt sind.

Reduktionen auf Gruppenebene

Die Bibliothek liefert fertige Kollektivoperationen. Eine Gruppen-reduce summiert eine Kachel mit einem einzigen Aufruf und verbirgt die Offset-Schleife, die Sie zuvor selbst geschrieben haben.

int total = cg::reduce(warp, val, cg::plus<int>());

Synchronisierbare Grids

Die leistungsfähigste Gruppe ist die Grid-Gruppe. Mit einem kooperativen Start kann jeder Block an einer Barriere synchronisieren – etwas, das ein normaler Kernel nicht kann.

Kooperativer Start erforderlich

Eine Synchronisierung über das gesamte Grid funktioniert nur, wenn Sie den Kernel mit cudaLaunchCooperativeKernel starten und die GPU dies unterstützt. Ein normaler Start erlaubt dies nicht.

Warum der Aufwand

Cooperative Groups machen Warp-Code lesbar und portierbar: keine manuell verwalteten Masken, klare Gültigkeitsbereiche und wiederverwendbare Kollektivoperationen, die Ihrer Absicht entsprechen.

Kurze Überprüfung

Rufen Sie sich ins Gedächtnis, wie Sie aus einem Block eine kooperative Gruppe in Warp-Größe erstellen.

Zusammenfassung

Cooperative Groups verwandeln Masken in benannte Objekte: Teilen Sie einen Block in Kacheln auf, rufen Sie reduce auf und synchronisieren Sie sogar ein gesamtes Grid. Sie beherrschen nun CUDA auf Warp-Ebene. ✨

Kostenlos starten

Lerne C++ mit einem KI-Tutor — kostenlos

Schreibe und führe echten Code in deinem Browser aus, bekomme sofortige Hilfe von einem 24/7 KI-Tutor und setze dein Lernen im Web oder in der App fort.

Kurse
30
Lektionen
120

Häufig gestellte Fragen

Ist die Lektion „Cooperative Groups“ kostenlos?

Ja — der vollständige Text von „Cooperative Groups“ ist hier im Web kostenlos zu lesen. Um sie interaktiv zu üben (integrierter Code-Editor und 24/7 KI-Tutor) und den Rest des CUDA Academy-Kurses freizuschalten, upgrade auf CoddyKit PRO. Der CUDA Academy-Kurs umfasst insgesamt 4 Lektionen.

Was lerne ich in „Cooperative Groups“?

Eine moderne API für flexible Synchronisierungsbereiche. Du übst CUDA Academy mit praktischem Code, den du direkt im Browser ausführst, und ein 24/7 KI-Tutor beantwortet deine Fragen während du die Lektion bearbeitest.

Brauche ich Erfahrung, um CUDA Academy zu starten?

Keine Vorkenntnisse erforderlich. CUDA Academy auf CoddyKit ist für Anfänger bis fortgeschrittene Lernende strukturiert, sodass du hier starten oder von Anfang an beginnen und in deinem eigenen Tempo voranschreiten kannst. Dies ist Lektion 4 von 4.

Wie lange dauert die Lektion „Cooperative Groups“?

Die meisten CoddyKit-Lektionen dauern etwa 5–10 Minuten. Jede ist kompakt und interaktiv, sodass du stetig Fortschritte machst und genau dort weitermachst, wo du aufgehört hast – im Web und in der App.

Kann ich in dieser CUDA Academy-Lektion Code schreiben und ausführen?

Ja. Jede CUDA Academy-Lektion enthält einen integrierten Code-Editor, sodass du echten Code direkt in deinem Browser schreibst und ausführst und sofort KI-Feedback erhältst — ohne lokale Einrichtung erforderlich.

Alle Lektionen in diesem Kurs

  1. Warps, Lanes und Masken
  2. __shfl_down_sync für Reduktionen
  3. Ballot- und Vote-Funktionen
  4. Cooperative Groups
← Zurück zu CUDA Academy