Ich habe zwei einfache Kernel codiert. Jeder fügt zwei Vektoren vom Typ int (32-Bit)/long int (64-Bit) hinzu. Es stellt sich heraus auf meiner GPU (Tesla K80), die ziemlich neu und gut ist, die Kerne sind nur 32-Bit.
Die Zeit verdoppelt sich ungefähr, wenn die Vektorgröße zunimmt.
Die Kerne sind wie folgt:
__global__ void add_32(int * c, int * a, int * b)
{
int gid = blockIdx.x * blockDim.x + threadIdx.x;
c[gid] = a[gid] + b[gid];
}
typedef long int int64;
__global__ void add_64(int64 * c, int64 * a, int64 * b)
{
int gid = blockIdx.x * blockDim.x + threadIdx.x;
c[gid] = a[gid] + b[gid];
}
Wenn Vektorgröße 1 Mega-Element, add_32 dauert etwa 102,911 microsec, während add_64 nimmt 192,669 microsec. (Ausführungszeiten wurden unter Verwendung des Nvidia-Profilers beim Ausführen der Binärdatei im Veröffentlichungsmodus gemeldet).
Es scheint, dass 64-Bit-Befehle nur über 32-Bit-Befehle emuliert werden!
Dies könnte eine Brute-Force-Lösung sein, um herauszufinden, welche Art von Maschinen die GPU-Kerne sind, aber definitiv keine elegante.
Update:
Dank @ Paul A. Clayton Kommentar, so scheint es, dass die Lösung über keinen fairen Vergleich ist die Datengröße in dem Fall 64-Bit verdoppelt. Daher sollten wir nicht beide Kernel mit der gleichen Anzahl von Elementen starten. Das korrekte Prinzip wäre, die 64-Bit-Version mit der halben Anzahl von Elementen zu starten.
Um noch sicherer zu sein, betrachten wir die elementweise Vektormultiplikation anstelle der Addition. Wenn die GPU 64-Bit-Befehle über 32-Bit-Befehle emuliert, benötigt sie mindestens 3 32-Bit-Multiplikationsbefehle, um 2 64-Bit-Zahlen zu multiplizieren, beispielsweise unter Verwendung des Karatsuba-Algorithmus. Dies impliziert, dass wenn wir den 64-Bit-Vektor-Multiplikationskern mit N/2 Elementen starten würden, würde es länger dauern als der 32-Bit-Kern mit N Elementen, wenn 64-Bit-Multiplikationen gerade emuliert würden.
Hier sind die Kerne:
__global__ void mul_32(int * c, int * a, int * b)
{
int gid = blockIdx.x * blockDim.x + threadIdx.x;
c[gid] = a[gid] * b[gid];
}
typedef long int int64;
__global__ void mul_64(int64 * c, int64 * a, int64 * b)
{
int gid = blockIdx.x * blockDim.x + threadIdx.x;
c[gid] = a[gid] * b[gid];
}
Und hier sind die Details Experiment: Zeiten hier gemeldet sind von nvidia-Profiler auf Release-Modus binary: 1- Kernel mul_32 mit Vektorgröße N = 256 Mega-Elemente , dauert 25.608 Millisekunden. 2- Kernel mul_64 mit Vektorgröße N = 128 Mega-Elemente, dauert 24.153 Millisekunden.
Ich bin mir bewusst, dass beide Kernel falsche Ergebnisse produzieren, aber ich denke, das hat nichts mit der Art und Weise zu tun, wie die Berechnung durchgeführt wird.
Ich bin nicht vertraut mit Nvidia GPUs, aber solche Informationen sollten in den Datenblättern oder anderen Handbüchern sein. Wenn es keine Informationen in der Öffentlichkeit gibt, benötigen Sie wahrscheinlich eine NDA mit Nvidia, um diese Informationen zu erhalten. Haben Sie Zugriff auf die Dokumentation der GPU, auf die Sie ausgerichtet sind? – fsasm
GPU-Spezifikationen zeigen diese Informationen nicht an. Ich denke, es sollte eine API/einen Befehl geben, der eine solche Information mitteilen kann !!! – caesar
Im Allgemeinen sollten Datenblätter solche Informationen anzeigen, weil dies ihr Zweck ist. Wenn der Anbieter die Informationen nicht veröffentlicht, brauchen Sie sie nicht. Der Treiber zusammen mit PTX verbergen alle Details der Hardware, um die Portabilität zu erhöhen. Wenn Sie diese Information wirklich brauchen, sollten Sie Nvidia kontaktieren. – fsasm