OpenCL - Speichergrößen

[Früchtemüsli]

Hi :)

Jetzt habe ich gerade etwas in der UNI-Vorlesung http://video2.virtuos.uni-osnabrueck.de:8080/engage/ui/watch.html?id=aa979fd7-9550-4cdb-8420-804d314b5ba5 gehört, was mich doch ziemlich verwirrt. Und zwar bei 08:45.

int index = get_group_id(0) * get_local_size(0) + get_local_id(0);


Also ich hatte das so in der vorhergehenden Vorlesung verstanden, dass man das selbst bestimmen kann, wie viele Work Items in Work Groups zusammengefasst werden. D. h. man fasst mehrere Hardware-Komponenten zu logischen Verbünden zusammen. Und zu jedem logischen Verbund (Work Group) gibt es dann einen zugehörigen lokalen Speicher.

Jetzt scheint es aber aufgrund dieser Anweisung so, dass die Größe dieser lokalen Speicher davon abhängt, wie groß der globale Speicher ist und wie viele Work Groups man definiert hat.

Das kommt mir irgendwie unlogisch vor. Warum kann man nicht mehr lokalen Speicher oder weniger lokalen Speicher definieren? Warum kann man den nicht beliebig groß definieren, also den jeweiligen Bedürfnissen anpassen?

[Gast]

Hi :)

Jetzt habe ich gerade etwas in der UNI-Vorlesung http://video2.virtuos.uni-osnabrueck.de:8080/engage/ui/watch.html?id=aa979fd7-9550-4cdb-8420-804d314b5ba5 gehört, was mich doch ziemlich verwirrt. Und zwar bei 08:45.




Also ich hatte das so in der vorhergehenden Vorlesung verstanden, dass man das selbst bestimmen kann, wie viele Work Items in Work Groups zusammengefasst werden.

Ja
D. h. man fasst mehrere Hardware-Komponenten zu logischen Verbünden zusammen. Und zu jedem logischen Verbund (Work Group) gibt es dann einen zugehörigen lokalen Speicher.

Was meinst du mit "Hardware Komponenten zusammenfassen"? Ein work item ist die kleinste Ausführungseinheit in OpenCL. In Cuda nennt man sie Threads. Das hat mit hardware nichts zu tun, ein work item wird von Hardware ausgeführt, so wie eine Funktion in C++ von der CPU ausgeführt wird. Du nennst deshalb aber auch nicht Funktionen in C++ plötzlich "Hardware Einheiten".

Jede work group hat lokalen Speicher, ja. Dessen Größe hängt von der Hardware ab, z.b. 48kB pro Core auf Fermi (Geforce 400), 64kB pro Core auf Kepler (Geforce 600)

Jetzt scheint es aber aufgrund dieser Anweisung so, dass die Größe dieser lokalen Speicher davon abhängt, wie groß der globale Speicher ist und wie viele Work Groups man definiert hat.

Wie kommst du darauf?
Du kannst soviel local memory verwenden wie du willst, nur ist es ev. nicht sehr sinnvoll weil der Vorteil von lokal memory die große Geschwindigkeit ist, es handelt sich ja um physikalischen Speicher direkt im Chip. Wenn du mehr verwendest als durch die Hardware vorgeben ist, muss in den global memory ausgewichen werden und der Geschwindigkeistvorteil ist dahin...

Das kommt mir irgendwie unlogisch vor. Warum kann man nicht mehr lokalen Speicher oder weniger lokalen Speicher definieren? Warum kann man den nicht beliebig groß definieren, also den jeweiligen Bedürfnissen anpassen?
Die Größe des local memory ist durch die Hardware vorgeben, da kannst du gar nichts definieren. Du kannst auch nicht den L1/L2 Cache von deiner CPU beliebig groß definieren und den Bedürfnissen anpassen... :|

[Gast]

Liegt vielleicht ein Missverstaendnis vor? get_local_size(0) und get_local_id(0) geben die Groesse einer Work-Group und die ID eines Work-Items innerhalb dieser Work-Group zurueck und haben nichts mit lokalen Speicher zu tun. Wie du auf lokalen Speicher kommst ist mir schleierhaft. get_group_id(0) gibt die ID der Work-Group zurueck.

Wieviel lokalen Speicher eine Work-Group (!) zugewiesen bekommt, kannst du selber dadurch bestimmen, dass ihn einfach deklarierst innerhalb einer Funktion, z.B. mit __local int array[256]. Dabei gibt es ein Maximum von einigen Kilobyte, abhaengig vom Device. Man sollte aber weit weg vom Maximum bleiben. Je mehr lokalen Speicher eine Work-Group nutzt, desto weniger Work-Groups koennen gleichzeitig einer Compute Unit zugewiesen werden, was schlecht ist.

[Früchtemüsli]

Heißt das jetzt, dass diese Berechnung in dieser Vorlesung eigentlich falsch ist?

Denn dann könnte es doch passieren, dass die Summe der lokalen Speicher größer ist als der globale Speicher, und der index dann ins Leere läuft, oder?

[Gast]

Heißt das jetzt, dass diese Berechnung in dieser Vorlesung eigentlich falsch ist?

Denn dann könnte es doch passieren, dass die Summe der lokalen Speicher größer ist als der globale Speicher, und der index dann ins Leere läuft, oder?
Lies die Antworten hier nochmal durch, die Berechnung

int index = get_group_id(0) * get_local_size(0) + get_local_id(0);

hat nichts mit der Groesse des local memory zu tun. Es geht hier um die eindeutige id eines work items, welche sich aus der grid/block struktur eben nach dieser Formel ergibt.