Kann die Summenberechnung in OpenCL parallel ausgeführt werden?

Question

Ich bin ein Neuling in OpenCL. Ich verstehe jedoch die C/C++ - Grundlagen und die OOP. Meine Frage lautet wie folgt: Ist es irgendwie möglich, die Summenberechnungsaufgabe parallel auszuführen? Ist es theoretisch möglich? Im Folgenden werde ich beschreiben, was ich zu tun versucht:

Die Aufgabe ist, zum Beispiel:

double* values = new double[1000]; //let's pretend it has some random values inside 
double sum = 0.0; 

for(int i = 0; i < 1000; i++) { 
    sum += values[i]; 
} 

Was ich in OpenCL-Kernel versucht zu tun (und ich glaube, es ist falsch, weil vielleicht greift er auf die gleiche "sum" -Variable aus verschiedenen Threads/Tasks gleichzeitig):

__kernel void calculate2dim(__global float* vectors1dim, 
          __global float output, 
          const unsigned int count) { 
    int i = get_global_id(0); 
    output += vectors1dim[i]; 
} 

Dieser Code ist falsch. Ich würde es sehr schätzen, wenn mir jemand antworten würde, wenn es theoretisch möglich wäre, solche Aufgaben parallel zu erledigen und wenn es - wie!

Answer 1

Wenn Sie die Werte Ihres Arrays parallel summieren möchten, sollten Sie sicherstellen, dass Sie Contention reduzieren und sicherstellen, dass es keine Datenabhängigkeiten über Threads gibt.

Datenabhängigkeiten verursachen, dass Threads auf einander warten müssen, wodurch Konflikte entstehen, die Sie vermeiden möchten, um eine echte Parallelisierung zu erhalten.

Eine Möglichkeit besteht darin, das Array in N-Arrays aufzuteilen, die jeweils einen Unterabschnitt des ursprünglichen Arrays enthalten, und dann die OpenCL-Kernelfunktion für jedes Array aufzurufen.

Am Ende, wenn alle Kernel die harte Arbeit getan haben, können Sie nur die Ergebnisse der einzelnen Arrays in eins zusammenfassen. Diese Operation kann leicht von der CPU durchgeführt werden.

Der Schlüssel ist, keine Abhängigkeiten zwischen den Berechnungen in jedem Kernel zu haben, also müssen Sie Ihre Daten und Verarbeitung entsprechend teilen.

Ich weiß nicht, ob Ihre Daten tatsächliche Abhängigkeiten von Ihrer Frage hat, aber das ist für Sie herauszufinden.

Answer 2

Das Stück Code, den ich als Referenz zur Verfügung gestellt habe, sollte die Arbeit machen.

z. Sie haben N Elemente, und die Größe Ihrer Arbeitsgruppe ist WS = 64. Ich nehme an, dass N ist ein Vielfaches von 2 * WS (das ist wichtig, berechnet eine Arbeitsgruppe Summe von 2 * WS-Elementen). Dann müssen Sie Kernel-Spezifizierungs auszuführen:

globalSizeX = 2*WS*(N/(2*WS)); 

Als Ergebnis Summe Array-Teilsummen 2 * WS Elemente haben. (z. B. Summe [1] - enthält die Summe der Elemente, deren Indizes von 2 * WS bis 4 * WS-1) sind.

Wenn Ihre globalSizeX ist 2 * WS oder weniger (was bedeutet, dass Sie nur eine Arbeitsgruppe haben), dann sind Sie fertig. Verwenden Sie einfach sum [0] als Ergebnis. Wenn nicht - müssen Sie den Vorgang wiederholen, diesmal mit Summe Array als Eingabe-Array und Ausgabe an andere Array (Erstellen Sie 2 Arrays und Ping-Pong zwischen ihnen). Und so weiter, bis Sie nur noch eine Arbeitsgruppe haben.

Suche auch nach Hilli Steele/Blelloch parallelen Algorithmen. This Artikel könnte nützlich sein, als auch

Hier ist das eigentliche Beispiel ist:

__kernel void par_sum(__global unsigned int* input, __global unsigned int* sum) 
{ 
    int li = get_local_id(0); 
    int groupId = get_group_id(0); 

    __local int our_h[2 * get_group_size(0)]; 
    our_h[2*li + 0] = hist[2*get_group_size(0)*blockId + 2*li + 0]; 
    our_h[2*li + 1] = hist[2*get_group_size(0)*blockId + 2*li + 1]; 

    // sweep up 
    int width = 2; 
    int num_el = 2*get_group_size(0)/width; 
    int wby2 = width>>1; 

    for(int i = 2*BLK_SIZ>>1; i>0; i>>=1) 
    { 

     barrier(CLK_LOCL_MEM_FENCE); 

     if(li < num_el) 
     { 
      int idx = width*(li+1) - 1; 
      our_h[idx] = our_h[idx] + our_h[(idx - wby2)]; 
     } 

     width<<=1; 
     wby2 = width>>1; 
     num_el>>=1; 
    } 

     barrier(CLK_LOCL_MEM_FENCE); 

    // down-sweep 
    if(0 == li) 
     sum[groupId] = our_h[2*get_group_size(0)-1]; // save sum 
}