CUDA 工程在 DGX Spark 上编译通过，启动内核时却出现 no kernel image is available for execution on the device，首先应检查产物覆盖的 GPU 架构，而不是重装整套环境。另一种更隐蔽的情况是程序一直依靠 PTX 即时编译运行，发布记录却误以为已经包含原生设备代码。本文用一个独立小程序，把目标查询、产物检查与两条执行路径串成验收流程。
一、分清主机架构与 GPU 目标
DGX Spark 使用 NVIDIA GB10 Grace Blackwell，主机架构为 ARM64，系统为 NVIDIA DGX OS，基于 Ubuntu 24.04 并预装 CUDA 栈。付款后在控制台查看连接凭据，经 SSH 操作即可。
ARM64 描述 CPU 侧可执行文件；CUDA 的计算能力描述 GPU 设备代码目标，两者不是同一个兼容性条件。file 显示 AArch64，只能说明主机侧格式正确，不能证明其中嵌入了可用内核。
SASS 是面向目标 GPU 的机器代码，PTX 是可由驱动即时编译的中间表示。发布包可以同时携带两者，但应分别检查，不能把普通运行成功当成两条路径都已验证。
以下实验只在新建目录中生成源码和可执行文件，不修改驱动、系统工具链或业务程序。示例结果只用于兼容性验收，不代表性能测试。
二、查询实际计算能力，不猜架构编号
先建立独立工作目录，保存系统与工具链信息。nvidia-smi 显示的 CUDA 版本反映驱动支持能力，不等于当前 nvcc 的工具包版本，因此两份输出都要留存。
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
将下面内容保存为 probe.cu。程序通过 CUDA Runtime 查询设备零的计算能力，失败时返回非零退出码，不会输出一个可误用的目标编号。
#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
这里故意不硬编码 GB10 的目标编号。若工具包列表不包含查询出的 sm_ 目标，应先确认工具包支持范围，不能随意换成较旧编号继续发布。探针失败则先处理设备访问或运行库问题，不进入编译阶段。
三、显式生成并检查 SASS 与 PTX
将下面内容保存为 check.cu。它只启动一个线程写入整数，再复制回主机核对结果。每个 CUDA 调用都检查错误，避免内核启动失败后仍打印成功。
#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;
}
在同一工作目录执行，显式要求同时生成原生代码与 PTX：
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
检查 ELF 列表及 SASS 输出中是否有对应目标和内核，PTX 输出是否包含 .version、.target 与内核入口。空文件、目标不符或缺少预期内核，都应阻断验收；仅检查命令退出码不够。
检查最终文件，而非中间对象
实际工程可能经过静态库链接、设备链接或打包步骤。检查对象文件中存在 PTX，并不能证明最终交付文件保留了它。应针对最终可执行文件及实际加载的共享库重复检查，并记录文件摘要。
四、分别验证原生执行与即时编译
先禁止 PTX 即时编译，要求当前产物依靠原生设备代码执行；再强制走 PTX 路径，并关闭磁盘编译缓存。两次测试均使用新进程，避免进程内已经加载的模块影响判断。
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
两份日志都应出现 PASS，退出码都应为零。测试变量仅作用于该次命令，不要写入全局环境，让业务进程长期强制 JIT。
原生失败而 PTX 成功，优先回查原生目标和最终链接；原生成功而 PTX 失败，检查 PTX 是否存在，以及驱动是否支持其版本。两者都失败，则结合具体错误检查工具链、运行库和设备状态，不以反复切换编译选项代替定位。
五、把实验结果变成发布门禁
小程序通过，只能证明基础编译链和这段内核在当前环境可用。业务还可能加载扩展库、按输入选择不同内核，或者在运行时生成代码，因此必须补充真实入口测试，覆盖主要数据类型与分支。
建议每次发布保留四组证据：
- 主机架构、驱动输出、工具包版本与实际编译命令。
- GPU 计算能力、最终文件的 SASS 与 PTX 检查结果。
- 两条执行路径的退出码、日志及业务结果核对记录。
- 源码版本、最终文件 SHA-256，以及动态依赖清单。
sha256sum ./check 可固定本次验收对象；后续重新链接，即使文件名不变，也需要重新验收。若使用容器，应在实际交付镜像内检查工具链及最终文件，在实际运行环境中完成执行验证。
PTX 是兼容性路径，不是对所有驱动的保证，更不能由一次成功推导出其他设备的运行结论。发布门禁应明确写成“当前驱动、当前产物、已覆盖入口通过”，架构或工具链发生变化后重新生成证据，而不是沿用旧日志。
常见问题
ARM64 可执行文件能启动，是否说明 CUDA 目标架构正确？
不能。ARM64 决定主机侧代码能否执行，GPU 设备代码还需覆盖实际计算能力。应检查嵌入的 SASS 与 PTX，并实际启动内核验收。
包含 PTX 是否保证所有驱动都能运行？
不保证。驱动必须支持产物使用的 PTX 版本及目标要求。强制 PTX 即时编译成功，只能证明当前驱动与当前产物的这条执行路径可用。
EPYC 7543P 独享物理主机，$472.5/月起