0
Код Snippit Calcs 3D-вектор величины: МАГ = SQRT (X * X + Y * Y + Z * Z)обмен данными Thread
// Note: blockDim.x = 300, gMem= ptr to global mem chunk
__shared__ sMem[100];
float regA;
for (j=0; j<50; j++) {
if(threadIdx.x < 3) {
regA= gMem[j];
sMem[threadIdx.x]= regA*regA; // Line A write 5193
}
__syncthreads();
if(threadIdx.x == 0) {
regA= sMem[0]; // Line B read 5197
regA+= sMem[1];
regA+= sMem[2]; // Line C read 5199
sMem[0]= sqrt(regA);
}
}
В syncthreads препятствующей RAW опасности. Без этого я получаю RAW & Предупреждения об опасности WAR. W/it, я все еще получаю предупреждения WAR. Предупреждение исходит от пар линий AB & AC.
WARN:(Warp Level Programming) Potential WAR hazard detected at __shared__ 0x30b in block (0, 0, 0) :
Read Thread (0, 0, 0) at 0x000000b0 in /src/trap.cu:5199:Mag(float const *, float const *, int, float*, int)
Write Thread (2, 0, 0) at 0x00000080 in /src/trap.cu:5193:Mag(float const *, float const *, int, float*, int)
Current Value : 64, Incoming Value : 66
WARN:(Warp Level Programming) Potential WAR hazard detected at __shared__ 0x307 in block (0, 0, 0) :
Read Thread (0, 0, 0) at 0x000000a8 in /src/trap.cu:5197::Mag(float const *, float const *, int, float*, int)
Write Thread (1, 0, 0) at 0x00000080 in /src/trap.cu:5193::Mag(float const *, float const *, int, float*, int)
Current Value : 67, Incoming Value : 66
Почему не __synthreads(); предотвратить опасения WAR?
Там нет промежуточных __synthreads(), когда код выполняет строки 5197 и 5199, а затем возвращается обратно line 5193. – njuffa
Да, в этом была проблема. Я пропустил тот факт, что строка 5197 выполняла строку B4 5193. (Ошибка пробела в заголовке оператора) – Doug
@njuffa, если вы разместите его в качестве ответа, я его выберу. –