Korrekte Art, OpenCL-Kernel vieler Speicherobjekte zu informieren?

Question

In meinem OpenCL-Programm werde ich mit mehr als 60 globalen Speicherpuffern enden, auf die jeder Kernel zugreifen müssen muss. Was ist die empfohlene Methode, um jedem Kernel die Position jedes dieser Puffer mitzuteilen?

Die Puffer selbst sind während der gesamten Lebensdauer der Anwendung stabil - das heißt, wir ordnen die Puffer beim Start der Anwendung zu, rufen mehrere Kernel auf und geben die Puffer nur auf der Anwendungsseite frei. Ihr Inhalt kann sich jedoch ändern, wenn die Kernel von ihnen lesen/schreiben.

In CUDA, die Art und Weise, die ich dies tat, war die Erstellung von mehr als 60 Programmumfang globale Variablen in meinem CUDA-Code. Ich würde dann auf dem Host die Adresse der Gerätepuffer schreiben, die ich diesen globalen Variablen zugewiesen habe. Dann würden Kernel diese globalen Variablen einfach verwenden, um den Puffer zu finden, mit dem sie arbeiten mussten.

Was wäre der beste Weg, dies in OpenCL zu tun? Es sieht so aus, als ob die globalen Variablen von CL ein bisschen anders sind als die von CUDA, aber ich kann keine klare Antwort darauf finden, ob meine CUDA-Methode funktionieren wird und wenn ja, wie die Pufferzeiger in globale Variablen übertragen werden. Wenn das nicht funktioniert, was ist sonst der beste Weg?

Answer 1

60 globale Variablen sind sicher eine Menge! Sind Sie sicher, dass es keinen Weg gibt, Ihren Algorithmus ein wenig zu refaktorisieren, um kleinere Datenblöcke zu verwenden? Denken Sie daran, jeder Kernel sollte eine minimale Arbeitseinheit sein, nicht etwas Kolossales!

Es gibt jedoch eine mögliche Lösung. Angenommen, Ihre 60 Arrays sind von bekannter Größe, können Sie sie alle in einem großen Puffer speichern und dann Offsets verwenden, um auf verschiedene Teile dieses großen Arrays zuzugreifen. Hier ist ein sehr einfaches Beispiel mit drei Arrays:

A is 100 elements 
B is 200 elements 
C is 100 elements 

big_array = A[0:100] B[0:200] C[0:100] 
offsets = [0, 100, 300] 

Dann müssen Sie nur big_array und Offsets zu den Kernel übergeben, und Sie können jedes Array zugreifen. Zum Beispiel:

A[50] = big_array[offsets[0] + 50] 
B[20] = big_array[offsets[1] + 20] 
C[0] = big_array[offsets[2] + 0] 

Ich bin nicht sicher, wie das Caching auf Ihrem spezielles Gerät beeinflussen würde, aber meine erste Vermutung ist, „nicht gut.“ Diese Art von Array-Zugriff ist auch ein wenig unangenehm. Ich bin mir nicht sicher, ob es gültig wäre, aber Sie könnten jeden Ihrer Kerne mit Code starten, der jeden Offset extrahiert und zu einer Kopie des ursprünglichen Zeigers hinzufügt.

Auf der Host-Seite können Sie clCreateSubBuffer: http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clCreateSubBuffer.html verwenden, damit Sie auch Referenzen auf bestimmte Arrays ohne das Offsets-Array übergeben können.

Ich denke nicht, dass diese Lösung besser als die Übergabe der 60 Kernel-Argumente ist, aber abhängig von Ihrer OpenCL-Implementierung clSetKernelArgs, könnte es schneller sein. Es wird sicherlich die Länge Ihrer Argumentliste reduzieren.

Answer 2

Sie müssen zwei Dinge tun. Zunächst wird jeder Kernel, das jeden globalen Speicherpuffer verwendet, sollte ein Argument für jedes ein, so etwas wie dies erklärt:

kernel void awesome_parallel_stuff(global float* buf1, ..., global float* buf60) 

, so dass jeder verwendeter Puffer für diesen Kernel aufgeführt ist. Und dann, auf der Host-Seite Seite, müssen Sie jeden Puffer erstellen und clSetKernelArg verwenden, um einen bestimmten Speicherpuffer an ein bestimmtes Kernel-Argument vor dem Aufruf clEnqueueNDRangeKernel an die Partei zu starten gestartet.

Beachten Sie, dass, wenn die Kernel den gleichen Puffer bei jeder Kernelausführung verwenden, Sie nur die Kernel-Argumente eine Zeit einrichten müssen. Ein häufiger Fehler, den ich sehen kann, der host-side Leistung bluten kann, ist, clSetKernelArg in Situationen wiederholt aufzurufen, in denen es völlig unnötig ist.