GPU-Programmierung

Beide Varianten (1) und (2) oben lassen eine Hälfte der Threads A und die andere Hälfte B ausführen. Welche Aussage zur Warp-Effizienz ist richtig (ein Warp habe 32 Threads)?

if (threadIdx.x % 2 == 0) A(); else B();   // (1)
if ((threadIdx.x / 32) % 2 == 0) A(); else B();   // (2)
  1. A(1) ist effizienter, da die Verzweigungen übereinstimmen
  2. BIn (1) teilt sich die Verzweigung in allen Warps auf, in (2) stimmt sie je Warp überein, daher ist (2) effizienter
  3. CBeide sind genau gleich effizient
Antwort und Erläuterung anzeigen

Antwort: B. In (1) teilt sich die Verzweigung in allen Warps auf, in (2) stimmt sie je Warp überein, daher ist (2) effizienter

In (1) wird nach geraden und ungeraden Threads aufgeteilt, sodass in jedem Warp Threads der A-Seite und der B-Seite gemischt sind und alle Warps A und B nacheinander ausführen. In (2) wird nach dem Wert der durch 32 geteilten Thread-Nummer aufgeteilt, sodass alle 32 Threads eines Warps auf derselben Seite liegen und jeder Warp nur A oder nur B ausführt. Auch bei gleicher Aufgabe verschwindet die Divergenz, wenn die Bedingung auf Warp-Einheiten abgestimmt wird.

Dieses Thema üben Zur Themenseite

Weitere Fragen zu GPU-Programmierung

  1. Welche Aussage zum Memory Coalescing der GPU ist richtig?
  2. Wird, wie oben gezeigt, mit der Thread-Nummer i ein Array in Abständen von 32 Elementen gelesen, ist dies langsamer als das Lesen von a[i]. Was ist der Hauptgrund?
  3. Welche Anordnung ist unter dem Gesichtspunkt des Memory Coalescing vorteilhaft, wenn Daten mit den Partikelpositionen x, y, z auf der GPU verarbeitet werden?
  4. Welche Aussage zum Shared Memory von CUDA ist richtig?
  5. Welche Funktion hat __syncthreads() in CUDA?

Vergleiche zu diesem Thema