Fabian
109 lines
3.7 KiB
C++
109 lines
3.7 KiB
C++
/**
|
|
* HIP Minimal Probe — Safe diagnostic for AMD BC-250
|
|
* Step-by-step: each step prints BEFORE attempting, so we know where it hangs/crashes
|
|
*/
|
|
#include <hip/hip_runtime.h>
|
|
#include <cstdio>
|
|
#include <unistd.h> // _exit() — bypasses C++ destructors
|
|
|
|
int main() {
|
|
printf("PROBE: Starting HIP minimal probe...\n");
|
|
fflush(stdout);
|
|
|
|
// Step 1: hipGetDeviceCount
|
|
printf("PROBE [1/6]: hipGetDeviceCount... ");
|
|
fflush(stdout);
|
|
int count = 0;
|
|
hipError_t err = hipGetDeviceCount(&count);
|
|
if (err != hipSuccess) {
|
|
printf("FAILED: %s\n", hipGetErrorString(err));
|
|
return 1;
|
|
}
|
|
printf("OK (%d devices)\n", count);
|
|
fflush(stdout);
|
|
|
|
if (count == 0) {
|
|
printf("PROBE: No devices. Exiting.\n");
|
|
return 1;
|
|
}
|
|
|
|
// Step 2: hipGetDeviceProperties
|
|
printf("PROBE [2/6]: hipGetDeviceProperties... ");
|
|
fflush(stdout);
|
|
hipDeviceProp_t props;
|
|
err = hipGetDeviceProperties(&props, 0);
|
|
if (err != hipSuccess) {
|
|
printf("FAILED: %s\n", hipGetErrorString(err));
|
|
return 1;
|
|
}
|
|
printf("OK\n");
|
|
printf(" Name: %s\n", props.name);
|
|
printf(" GCN Arch: %s\n", props.gcnArchName);
|
|
printf(" CUs: %d\n", props.multiProcessorCount);
|
|
printf(" Total Mem: %zu MB\n", props.totalGlobalMem / (1024*1024));
|
|
printf(" Managed Memory: %s\n", props.managedMemory ? "YES" : "NO");
|
|
printf(" Concurrent Managed: %s\n", props.concurrentManagedAccess ? "YES" : "NO");
|
|
printf(" Integrated (APU): %s\n", props.integrated ? "YES" : "NO");
|
|
printf(" pageableMemoryAccess: %s\n", props.pageableMemoryAccess ? "YES" : "NO");
|
|
fflush(stdout);
|
|
|
|
// Step 3: hipSetDevice
|
|
printf("PROBE [3/6]: hipSetDevice(0)... ");
|
|
fflush(stdout);
|
|
err = hipSetDevice(0);
|
|
if (err != hipSuccess) {
|
|
printf("FAILED: %s\n", hipGetErrorString(err));
|
|
return 1;
|
|
}
|
|
printf("OK\n");
|
|
fflush(stdout);
|
|
|
|
// Step 4: Try hipHostMalloc (pinned host memory — safest for APU)
|
|
printf("PROBE [4/6]: hipHostMalloc (64 KB, coherent)... ");
|
|
fflush(stdout);
|
|
float* hostPtr = nullptr;
|
|
err = hipHostMalloc(&hostPtr, 64 * 1024, hipHostMallocCoherent);
|
|
if (err != hipSuccess) {
|
|
printf("FAILED: %s\n", hipGetErrorString(err));
|
|
printf(" Trying hipHostMallocDefault...\n");
|
|
err = hipHostMalloc(&hostPtr, 64 * 1024, hipHostMallocDefault);
|
|
if (err != hipSuccess) {
|
|
printf(" Also FAILED: %s\n", hipGetErrorString(err));
|
|
return 1;
|
|
}
|
|
}
|
|
printf("OK (ptr=%p)\n", (void*)hostPtr);
|
|
fflush(stdout);
|
|
|
|
// Step 5: Try hipMallocManaged (unified memory)
|
|
printf("PROBE [5/6]: hipMallocManaged (64 KB)... ");
|
|
fflush(stdout);
|
|
float* managedPtr = nullptr;
|
|
err = hipMallocManaged(&managedPtr, 64 * 1024);
|
|
if (err != hipSuccess) {
|
|
printf("FAILED: %s (this may be expected without XNACK)\n", hipGetErrorString(err));
|
|
fflush(stdout);
|
|
} else {
|
|
printf("OK (ptr=%p)\n", (void*)managedPtr);
|
|
fflush(stdout);
|
|
// Write test
|
|
managedPtr[0] = 42.0f;
|
|
printf(" Write test: managedPtr[0] = %f\n", managedPtr[0]);
|
|
fflush(stdout);
|
|
hipFree(managedPtr);
|
|
}
|
|
|
|
// Step 6: Free host memory — do NOT call hipDeviceReset()!
|
|
// hipDeviceReset() causes KIQ fence timeout → system hang on BC-250
|
|
printf("PROBE [6/6]: Cleanup (no device reset)... ");
|
|
fflush(stdout);
|
|
hipHostFree(hostPtr);
|
|
// hipDeviceReset() intentionally omitted — crashes BC-250
|
|
printf("OK\n");
|
|
fflush(stdout);
|
|
|
|
printf("\nPROBE: ALL STEPS COMPLETED SUCCESSFULLY\n");
|
|
fflush(stdout);
|
|
_exit(0); // HARD EXIT — bypasses HIP runtime destructors that crash BC-250
|
|
}
|