Wenn eindimensionale Gitter zu klein sind, verwenden Sie stattdessen zweidimensionale (oder dreidimensionale auf Fermi mit CUDA 4.0) Gitter. Die Dimensionalität in Raster- und Blocklayouts dient eigentlich nur der Bequemlichkeit - sie lässt den Ausführungsraum wie die üblichen parallelen Dateneingabebereiche aussehen, mit denen Programmierer zu arbeiten gewohnt sind (Matrizen, Raster, Voxel usw.). Aber es ist nur eine sehr kleine Abstraktion vom zugrundeliegenden einfachen linearen Nummerierungsschema, das über 10^12 eindeutige Thread-IDs innerhalb eines einzigen Kernel-Starts handhaben kann.
Bei Gittern ist die Reihenfolge spaltenbezogen. Wenn Sie also zuvor ein 1D-Gitterproblem hatten, wurde der "eindeutige 1D-Fadenindex" wie folgt berechnet:
unsigned int tid = threadIdx.x + blockIdx.x * blockDim.x;
was eine theoretische Obergrenze von 512 * 65535 = 33553920 einzelnen Threads hat. Das entsprechende 2D-Gitterproblem ist nur eine einfache Erweiterung des 1D-Falls
size_t tidx = threadIdx.x + blockIdx.x * blockDim.x;
size_t tid = tidx + blockIdx.y * blockDim.x * GridDim.x;
das eine theoretische Obergrenze von 512 * 65535 * 65535 = 2198956147200 einzelnen Threads hat. Fermi erlaubt es, dem Gitter eine dritte Dimension hinzuzufügen, ebenfalls mit einer maximalen Größe von 65535, was bis zu etwa 10^17 Threads in einem einzigen Ausführungsgitter ergibt. Das ist ziemlich viel.
