CUDA-Artefakte auf DGX Spark prüfen: Zielarchitektur, SASS und PTX

Ein CUDA-Projekt lässt sich auf DGX Spark erfolgreich kompilieren, doch beim Start eines Kernels erscheint no kernel image is available for execution on the device. In diesem Fall sollten Sie zuerst prüfen, welche GPU-Architekturen das Build-Artefakt abdeckt, statt die gesamte Umgebung neu zu installieren. Weniger offensichtlich ist ein anderer Fall: Das Programm läuft schon die ganze Zeit über die PTX-Just-in-Time-Kompilierung, während die Release-Dokumentation fälschlich davon ausgeht, dass nativer Gerätecode enthalten ist. Dieser Artikel verbindet die Ermittlung der Zielarchitektur, die Prüfung der Artefakte und den Test beider Ausführungspfade anhand eines eigenständigen kleinen Programms zu einem Abnahmeverfahren.

1. Host-Architektur und GPU-Zielarchitektur unterscheiden

DGX Spark verwendet NVIDIA GB10 Grace Blackwell mit ARM64 als Host-Architektur. Als Betriebssystem kommt NVIDIA DGX OS auf Basis von Ubuntu 24.04 mit vorinstalliertem CUDA-Stack zum Einsatz. Nach der Zahlung können Sie die Zugangsdaten in der Konsole abrufen und das System per SSH bedienen.

ARM64 beschreibt die Architektur der CPU-seitigen ausführbaren Datei; die CUDA Compute Capability beschreibt dagegen das Ziel für den GPU-Gerätecode. Es handelt sich um zwei unterschiedliche Kompatibilitätsbedingungen. Wenn file AArch64 ausgibt, bestätigt das lediglich das korrekte Format auf der Host-Seite. Es belegt nicht, dass nutzbare Kernel eingebettet sind.

SASS ist Maschinencode für die Ziel-GPU, PTX eine Zwischendarstellung, die der Treiber zur Laufzeit kompilieren kann. Ein Release-Paket kann beides enthalten. Beide Bestandteile müssen jedoch getrennt geprüft werden: Ein erfolgreicher normaler Programmlauf bestätigt nicht automatisch beide Pfade.

Die folgenden Versuche erzeugen Quellcode und ausführbare Dateien ausschließlich in einem neu angelegten Verzeichnis. Treiber, System-Toolchain und Anwendungsprogramme bleiben unverändert. Die Beispielergebnisse dienen nur der Kompatibilitätsprüfung, nicht als Leistungstest.

2. Die tatsächliche Compute Capability abfragen, statt die Architekturkennung zu erraten

Legen Sie zunächst ein separates Arbeitsverzeichnis an und sichern Sie die Angaben zu System und Toolchain. Die von nvidia-smi angezeigte CUDA-Version gibt den vom Treiber unterstützten Stand an. Sie ist nicht mit der Toolkit-Version des aktuell verwendeten nvcc gleichzusetzen. Bewahren Sie deshalb beide Ausgaben auf.

mkdir -p "$HOME/cuda-artifact-check"
cd "$HOME/cuda-artifact-check"
uname -m > host-arch.txt
nvidia-smi > driver.txt
nvcc --version > toolkit.txt
nvcc --list-gpu-code > supported-targets.txt

Speichern Sie den folgenden Inhalt als probe.cu. Das Programm fragt über die CUDA Runtime die Compute Capability von Gerät null ab. Schlägt die Abfrage fehl, liefert es einen Exit-Code ungleich null und gibt keine Zielkennung aus, die versehentlich weiterverwendet werden könnte.

#include <cuda_runtime.h>
#include <cstdio>

int main() {
    cudaDeviceProp p{};
    cudaError_t e = cudaGetDeviceProperties(&p, 0);
    if (e != cudaSuccess) {
        std::fprintf(stderr, "%s\n", cudaGetErrorString(e));
        return 1;
    }
    std::printf("%d%d\n", p.major, p.minor);
    return 0;
}

nvcc probe.cu -o probe
CC=$(./probe) || exit 1
printf '%s\n' "$CC" > compute-capability.txt
grep -Fx "sm_${CC}" supported-targets.txt || exit 1

Die Zielkennung für GB10 wird hier bewusst nicht fest einprogrammiert. Enthält die Toolkit-Liste das ermittelte sm_-Ziel nicht, klären Sie zunächst, welche Ziele das Toolkit unterstützt. Ersetzen Sie die Kennung nicht einfach durch eine ältere, um das Release trotzdem fortzusetzen. Schlägt das Prüfprogramm fehl, beheben Sie zuerst Probleme beim Gerätezugriff oder mit den Laufzeitbibliotheken, bevor Sie mit der Kompilierung fortfahren.

3. SASS und PTX explizit erzeugen und prüfen

Speichern Sie den folgenden Inhalt als check.cu. Das Programm startet lediglich einen Thread, der einen Integer-Wert schreibt, und kopiert das Ergebnis zur Prüfung zurück auf den Host. Jeder CUDA-Aufruf wird auf Fehler geprüft, damit ein fehlgeschlagener Kernel-Start nicht trotzdem als Erfolg ausgegeben wird.

#include <cuda_runtime.h>
#include <cstdio>

#define CUDA_OK(call) do { \
    cudaError_t e = (call); \
    if (e != cudaSuccess) { \
        std::fprintf(stderr, "%s\n", cudaGetErrorString(e)); \
        return 1; \
    } \
} while (0)

__global__ void write_value(int* p) {
    *p = 7;
}

int main() {
    int* device = nullptr;
    int host = 0;
    CUDA_OK(cudaMalloc(&device, sizeof(int)));
    write_value<<<1, 1>>>(device);
    CUDA_OK(cudaGetLastError());
    CUDA_OK(cudaDeviceSynchronize());
    CUDA_OK(cudaMemcpy(&host, device, sizeof(int),
                       cudaMemcpyDeviceToHost));
    CUDA_OK(cudaFree(device));
    if (host != 7) return 2;
    std::puts("PASS");
    return 0;
}

Führen Sie die folgenden Befehle im selben Arbeitsverzeichnis aus. Dabei wird ausdrücklich festgelegt, dass sowohl nativer Code als auch PTX erzeugt werden:

CC=$(cat compute-capability.txt)
nvcc -O2 \
  -gencode "arch=compute_${CC},code=[sm_${CC},compute_${CC}]" \
  check.cu -o check

file check > binary-format.txt
cuobjdump --list-elf ./check > elf-list.txt
cuobjdump --dump-sass ./check > sass.txt
cuobjdump --dump-ptx ./check > ptx.txt

Prüfen Sie, ob die ELF-Liste und die SASS-Ausgabe das passende Ziel und den erwarteten Kernel enthalten. Die PTX-Ausgabe muss .version, .target und den Kernel-Einstiegspunkt enthalten. Leere Dateien, ein falsches Ziel oder ein fehlender erwarteter Kernel müssen die Abnahme verhindern. Nur den Exit-Code der Befehle zu prüfen, reicht nicht aus.

Die endgültige Datei prüfen, nicht die Zwischenergebnisse

In realen Projekten können das Linken statischer Bibliotheken, Device Linking oder Paketierungsschritte hinzukommen. Dass eine Objektdatei PTX enthält, beweist nicht, dass es auch in der endgültig ausgelieferten Datei erhalten bleibt. Wiederholen Sie die Prüfung für die finale ausführbare Datei und die tatsächlich geladenen Shared Libraries und dokumentieren Sie die Datei-Hashwerte.

4. Native Ausführung und Just-in-Time-Kompilierung getrennt validieren

Deaktivieren Sie zunächst die PTX-Just-in-Time-Kompilierung, sodass das aktuelle Artefakt auf nativen Gerätecode angewiesen ist. Erzwingen Sie anschließend den PTX-Pfad und deaktivieren Sie dabei den Kompilierungs-Cache auf dem Datenträger. Beide Tests laufen jeweils in einem neuen Prozess, damit bereits im Prozess geladene Module das Ergebnis nicht beeinflussen.

env -u CUDA_FORCE_PTX_JIT \
  CUDA_DISABLE_PTX_JIT=1 \
  ./check > native.log 2>&1
native_rc=$?

env -u CUDA_DISABLE_PTX_JIT \
  CUDA_FORCE_PTX_JIT=1 CUDA_CACHE_DISABLE=1 \
  ./check > ptx-jit.log 2>&1
ptx_rc=$?

printf 'native=%s\nptx_jit=%s\n' \
  "$native_rc" "$ptx_rc" > execution-status.txt

cat execution-status.txt native.log ptx-jit.log
test "$native_rc" -eq 0 && test "$ptx_rc" -eq 0

Beide Protokolle müssen PASS enthalten, und beide Exit-Codes müssen null sein. Die Testvariablen gelten nur für den jeweiligen Befehl. Tragen Sie sie nicht in die globale Umgebung ein, da dies für Anwendungsprozesse dauerhaft JIT erzwingen würde.

Scheitert die native Ausführung, während PTX funktioniert, prüfen Sie zuerst das native Ziel und den abschließenden Link-Schritt. Funktioniert die native Ausführung, während PTX scheitert, prüfen Sie, ob PTX vorhanden ist und ob der Treiber dessen Version unterstützt. Scheitern beide Pfade, untersuchen Sie anhand der konkreten Fehlermeldungen die Toolchain, die Laufzeitbibliotheken und den Gerätezustand. Wiederholtes Umschalten von Compiler-Optionen ersetzt keine Fehleranalyse.

5. Testergebnisse als Freigabekriterium für Releases nutzen

Ein erfolgreicher Test des kleinen Programms belegt lediglich, dass die grundlegende Build-Toolchain und dieser Kernel in der aktuellen Umgebung funktionieren. Die Anwendung kann zusätzliche Bibliotheken laden, abhängig von den Eingaben andere Kernel auswählen oder zur Laufzeit Code erzeugen. Deshalb sind ergänzende Tests der tatsächlichen Anwendungseinstiegspunkte erforderlich, die die wichtigsten Datentypen und Verzweigungen abdecken.

Bewahren Sie für jedes Release vier Gruppen von Nachweisen auf:

• Host-Architektur, Treiberausgabe, Toolkit-Version und den tatsächlich verwendeten Kompilierungsbefehl.
• Compute Capability der GPU sowie die Ergebnisse der SASS- und PTX-Prüfung der endgültigen Dateien.
• Exit-Codes und Protokolle beider Ausführungspfade sowie Aufzeichnungen zur Überprüfung der Anwendungsergebnisse.
• Quellcodeversion, SHA-256 der endgültigen Dateien und eine Liste der dynamischen Abhängigkeiten.

Mit sha256sum ./check lässt sich das geprüfte Artefakt eindeutig festhalten. Wird es später erneut gelinkt, ist auch bei unverändertem Dateinamen eine neue Abnahme erforderlich. Wenn Sie Container verwenden, prüfen Sie Toolchain und endgültige Dateien im tatsächlich ausgelieferten Image und führen Sie die Ausführungstests in der tatsächlichen Laufzeitumgebung durch.

PTX ist ein Kompatibilitätspfad, keine Garantie für die Unterstützung durch jeden Treiber. Aus einem einzelnen erfolgreichen Test lässt sich erst recht nicht ableiten, dass das Programm auch auf anderen Geräten läuft. Das Freigabekriterium sollte ausdrücklich lauten: „Mit dem aktuellen Treiber und dem aktuellen Artefakt sind die abgedeckten Einstiegspunkte erfolgreich geprüft.“ Ändern sich Architektur oder Toolchain, müssen neue Nachweise erzeugt werden, statt alte Protokolle weiterzuverwenden.

Häufig gestellte Fragen