CUDA Academy · Lektion

Ballot- und Vote-Funktionen

Prüfen Sie Prädikate über einen Warp hinweg.

Lektion 3 von 413 Schritte

Ballot- und Vote-Funktionen ist eine kostenlose CUDA Academy-Lektion auf CoddyKit. Dies ist Lektion 3 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.

Lanes können abstimmen

Neben dem Verschieben von Daten kann ein Warp auch gemeinsam Ja-Nein-Fragen beantworten. Mit Vote-Intrinsics kann jede Lane ihr Wahr-oder-Falsch-Prädikat in einem einzigen kostengünstigen Schritt weitergeben.

Lernen Sie ballot_sync kennen

Die flexibelste Abstimmungsfunktion ist ballot_sync. Sie gibt eine 32-Bit-Zahl zurück, bei der Bit i genau dann 1 ist, wenn das Prädikat von Lane i wahr war.

unsigned bits = __ballot_sync(mask, pred);

Ein Bit pro Lane

Die zurückgegebene Ganzzahl ist eine kleine Bitmap des Warps. Das Lesen von Bit i zeigt, ob Lane i mit „wahr“ abgestimmt hat – alles in einem einzigen Register.

Die Wahr-Werte zählen

Sie möchten wissen, wie viele Lanes mit Ja gestimmt haben? Zählen Sie die gesetzten Bits des Abstimmungsergebnisses. popc zählt die gesetzten Bits und liefert sofort die Anzahl der Wahr-Stimmen im Warp.

int yes = __popc(__ballot_sync(mask, pred));

Alle oder irgendeine

Zwei Abkürzungen beantworten häufige Fragen. all_sync gibt nur dann true zurück, wenn jede Lane zugestimmt hat, während any_sync true zurückgibt, wenn mindestens eine Lane zugestimmt hat.

bool every = __all_sync(mask, pred);

Any bedeutet mindestens eine

Verwenden Sie any_sync für vorzeitige Abbrüche: Wenn eine beliebige Lane eine Übereinstimmung gefunden hat, kann der gesamte Warp gemeinsam reagieren, ohne Lane für Lane zu durchsuchen.

bool found = __any_sync(mask, pred);

Stream-Kompaktierung

Ballot ermöglicht Kompaktierung: Jede Lane entscheidet, ob sie ihr Element behält, stimmt ab, und die Bitmap zeigt den Lanes, wohin sie in ein gepacktes Ausgabe-Array schreiben sollen.

Den eigenen Platz finden

Um ihren Wert zu platzieren, zählt eine Lane die gesetzten Wahr-Bits unterhalb ihrer eigenen Position. Dieser Präfixzähler aus dem Ballot ist ihr Index im kompaktierten Ergebnis.

int slot = __popc(bits & ((1u << lane) - 1));

Eine führende Lane auswählen

Manchmal soll eine Lane eine gemeinsame Aufgabe übernehmen. Das niedrigste gesetzte Bit im Ballot, das mit ffs gefunden wird, bezeichnet eine natürliche führende Lane, auf die sich der gesamte Warp einigen kann.

Immer eine Maske übergeben

Wie Shuffles verwenden auch Vote-Intrinsics eine aktive Maske. Inaktive Lanes tragen ein Bit mit dem Wert 0 bei. Übergeben Sie daher die korrekte Menge, damit keine nicht vorhandenen Lanes gezählt werden.

Kostengünstige Koordination

Abstimmungen kosten nur eine Instruktion und greifen nicht auf den Speicher zu. Sie ermöglichen es einem Warp, Entscheidungen deutlich günstiger zu koordinieren, als dies mit Atomics oder Flags im Shared Memory möglich wäre.

Kurze Überprüfung

Rufen Sie sich ins Gedächtnis, was ballot_sync an jede Lane zurückgibt.

Zusammenfassung

Sie haben einen Warp mit ballot und Vote abgefragt: Zählen Sie Wahr-Werte, prüfen Sie, ob alle oder irgendeine Lane zugestimmt hat, und kompaktieren Sie Datenströme kostengünstig. Als Nächstes: Cooperative Groups für flexible Synchronisierung. ✨

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 „Ballot- und Vote-Funktionen“ kostenlos?

Ja — der vollständige Text von „Ballot- und Vote-Funktionen“ 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 „Ballot- und Vote-Funktionen“?

Prüfen Sie Prädikate über einen Warp hinweg. 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 3 von 4.

Wie lange dauert die Lektion „Ballot- und Vote-Funktionen“?

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