CUDA Shared-Memory-Array - ungeradees Verhalten

Question

In einem CUDA Kernel, ich habe Code ähnlich dem folgenden. Ich versuche einen Zähler pro Thread zu berechnen, und akkumulieren die Zähler über den Block einen Nenner zu berechnen und dann das Verhältnis zurückzukehren. Jedoch ist CUDA den Wert denom Einstellung auf den Wert, für numer durch das Gewinde in dem Block mit der größten threadIdx.x, eher als die Summe des numer Wertes berechnet über alle Threads in dem Block berechnet. Wer weiß, was los ist?

extern __shared__ float s_shared[];

float numer = //calculate numerator

s_shared[threadIdx.x] = numer;
s_shared[blockDim.x] += numer;
__syncthreads();

float denom = s_shared[blockDim.x];
float result = numer/denom;

„Ergebnis“ sollte immer zwischen 0 und 1 und ist auf 1 über den Block zusammenzufassen, sondern es ist gleich 1,0 für jeden Thread, wo threadIdx.x das Maximum ist, und ein anderer Wert nicht auf den Bereich beschränkt für die anderen Threads in dem Block.

Answer 1

Sie Synchronisierung nicht die Summierung richtig an die blockDim.x Lage. Keiner der Threads warten, um zu sehen, was andere geschrieben haben, bevor ihre Summe addieren. So ähnlich wie

• Jeder liest Null,
• geht nach Hause, berechnet Null + numer.
• Everone schreibt Null + numer auf den Speicherplatz

Die hohe ThreadID gewinnt b / c es eine hohe Wahrscheinlichkeit handelt zuletzt hat, nehme ich an.

Was Sie, um stattdessen tun wollen eine schnelle Summe zu tun, ist eine binäre Summe zu tun auf s_shared[threadIdx.x]

• jeder schreibt ihre numer
• die Hälfte der Fäden berechnen Summen von Paaren und diejenigen, an einen neuen Standort schreiben
• ein Viertel der Fäden caluclate die Summen von Paaren von Paaren, und diejenigen, an einen neuen Standort schreiben
• etc
• , bis Sie haben nur einen Thread und eine Summe

Dieser Vorgang dauert O (n) Arbeit und O-Zeit (n log).