Divergenz in CUDA - Beenden eines Threads im Kernel

Question

Ich frage mich, wie ich aus einem Thread, dessen Thread-Index zu groß ist, beenden kann. Ich sehe zwei Möglichkeiten:

int i = threadIdx.x; 
if(i >= count) 
    return; 
// do logic 

oder

int i = threadIdx.x; 
if(i < count) { 
    // do logic 
} 

Ich weiß, dass beide korrekt sind, aber welche mehr die Leistung auswirken?

Answer 1

Obwohl beide in Bezug auf die Leistung gleich sind, sollten Sie berücksichtigen, dass die erste nicht empfohlen wird.

Die Rückgabe eines Threads innerhalb eines Kernels kann zu einem unerwarteten Verhalten im Rest des Codes führen.

Mit unerwartetem Verhalten meine ich jedes Problem im Zusammenhang mit der minimalen Einheit von Threads, die in einem Warp gruppiert sind. In einem Beispiel, wenn Sie einen Block in Ihrem Kernel haben, wird diese Situation als Thread-Divergenz bezeichnet und führt im Normalfall dazu, dass Threads im Leerlauf bleiben und andere Anweisungen ausführen.

CUDA von Beispiel Buch, Kapitel 5, Thema Kooperation:

Aber im Fall von __syncthreads(), ist das Ergebnis ein wenig tragisch. Die CUDA-Architektur gewährleistet, dass kein Thread über die __syncthreads auf einen Befehl vorrücken wird(), bis jeder Faden in dem Block die __syncthreads ausgeführt hat()

Also, es vor allem auf die Themen-Synchronisation innerhalb eines Kernel verwendet ist. Sie können eine sehr gute Frage/Antwort zu diesem Thema finden Sie hier: Can I use __syncthreads() after having dropped threads?

Wie ich letzte Anmerkung, ich habe auch verwendet, dass schlechte Praxis und kein Problem erschien, aber es gibt keine Garantie, dass Probleme in der Zukunft entstehen können . Es ist etwas, was ich nicht empfehlen würde