CUDA Koordinaten Treffer

Question

Ich versuche eine ziemlich einfache Mittelung während der Transformation eines Bildes zu implementieren. Ich habe die Transformation bereits erfolgreich implementiert, aber jetzt muss ich das resultierende Bild verarbeiten, indem ich alle Pixel aller 5x5 Pixel Rechtecke summiere. Meine Idee war, einen Zähler für jeden solchen 5x5-Block zu erhöhen, wenn ein Pixel in diesem Block gesetzt wird. Diese Blockzähler werden jedoch bei weitem nicht oft genug inkrementiert. So zum Debuggen überprüfte ich, wie oft jedes Pixel eines solchen Blocks überhaupt getroffen wird:

int x = (blockIdx.x*blockDim.x) + threadIdx.x; 
    int y = (blockIdx.y*blockDim.y) + threadIdx.y; 

    if((x<5)&&(y<5)) 
{ 
    resultArray [0]++; 
} 

Der Kernel wie folgt aufgerufen wird:

dim3 threadsPerBlock(8, 8); 
dim3 grid(targetAreaRect_px._uiWidth/threadsPerBlock.x, targetAreaRect_px._uiHeight/threadsPerBlock.y); 
CudaTransformAndAverageImage << < grid, threadsPerBlock >> > (pcPreRasteredImage_dyn, resultArray); 

Ich würde erwarten, resultArray [0] 25 enthalten nach Kernel-Ausführung, aber es enthält nur 1. Ist dies aufgrund einiger Optimierung durch den CUDA-Compiler?

Answer 1

Dies:

if((x<5)&&(y<5)) 
{ 
    resultArray [0]++; 
} 

ist ein Lesen nach dem Schreiben Gefahr.

Alle Threads, die (x<5)&&(y<5) erfüllen, können möglicherweise gleichzeitige Lese- und Schreibvorgänge von resultArray[0] versuchen. Das CUDA-Ausführungsmodell garantiert nicht die Reihenfolge der gleichzeitigen Speichertransaktionen.

Sie könnten diese Arbeit durch Atom Speicher-Transaktionen unter Verwendung, zum Beispiel:

if((x<5)&&(y<5)) { 
    atomicAdd(&resultArray[0], 1); 
} 

Dadurch wird die Speichertransaktionen serialisiert werden und die Berechnung korrekt machen. Dies wird sich auch sehr negativ auf die Leistung auswirken.

Sie könnten untersuchen, ob jeder Block eine lokale Summe mit einer Reduktionstypberechnung berechnet und dann die lokalen Blocksummen atomar oder auf dem Host oder in einem zweiten Kernel summiert.