Problem Description
Running the folloing coed produces the error:
AMDGPU message: Running a program that requires XNACK on a system where XNACK is disabled. This may cause problems when using an OS-allocated pointer inside a target region. Re-run with HSA_XNACK=1 to remove this warning.
Display only launched kernel:
Kernel 'omp target in step_physics_usm @ 38 (__omp_offloading_10302_2ba2398_step_physics_usm_l38)'
OFFLOAD ERROR: memory access fault by GPU 1 (agent 0x18415920) at virtual address 0x7fffa7d19000. Reasons: Page not present or supervisor privilege, Write access to a read-only page
Use 'OFFLOAD_TRACK_ALLOCATION_TRACES=true' to track device allocations
Aborted (core dumped)
The error still occurs when I set ```export HSA_XNACK=1`` on the command line and in my .bashrc file.
I am compiling with:
amdclang -O3 -target x86_64-linux-gnu -fopenmp --offload-arch=gfx1151 test_gpu_offload_unified_sm.c
#include <omp.h>
#include <stdio.h>
#pragma omp requires unified_shared_memory
void step_physics_usm(int n, float* pos, float* vel, float dt, int* ran_on_device) {
int device_flag = 0;
#pragma omp target map(tofrom: device_flag)
{
if (!omp_is_initial_device()) {
device_flag = 1;
}
}
#pragma omp target teams distribute parallel for //simd
for(int i = 0; i < n; i++) {
pos[i] += vel[i] * dt;
}
*ran_on_device = device_flag;
}
int n = 100000;
int main()
{
float pos[n];
float vel[n];
float dt = 0.016;
for (int i = 0; i < n; i++) {
pos[i] = 0.0;
vel[i] = 10.0;
}
int ranOnDevice = 0;
printf("Sending computation to OpenMP...");
// Pass the raw memory pointers to the C function
step_physics_usm(n, &pos[0], &vel[0], dt, &ranOnDevice);
// Verify the results
if (ranOnDevice == 1)
printf("[SUCCESS] Computation successfully offloaded to the APU!");
else {
printf("[WARNING] Computation fell back to the Host CPU.\n");
printf(" Check your compiler offload flags, OpenMP installation, and GPU drivers.\n");
}
printf("Sample result: pos[0] = %f\n", pos[0]);
return 0;
}
### Operating System
Ubuntu 25.10
### CPU
AMD RYZEN AI MAX+ 395
### GPU
AMD Radeon 8060S
### ROCm Version
7.2
### ROCm Component
aomp
### Steps to Reproduce
_No response_
### (Optional for Linux users) Output of /opt/rocm/bin/rocminfo --support
_No response_
### Additional Information
_No response_
Problem Description
Running the folloing coed produces the error:
The error still occurs when I set ```export HSA_XNACK=1`` on the command line and in my .bashrc file.
I am compiling with:
#include <omp.h>
#include <stdio.h>
#pragma omp requires unified_shared_memory
void step_physics_usm(int n, float* pos, float* vel, float dt, int* ran_on_device) {
int device_flag = 0;
#pragma omp target map(tofrom: device_flag)
{
if (!omp_is_initial_device()) {
device_flag = 1;
}
}
#pragma omp target teams distribute parallel for //simd
for(int i = 0; i < n; i++) {
pos[i] += vel[i] * dt;
}
*ran_on_device = device_flag;
}
int n = 100000;
int main()
{
float pos[n];
float vel[n];
float dt = 0.016;
for (int i = 0; i < n; i++) {
pos[i] = 0.0;
vel[i] = 10.0;
}
int ranOnDevice = 0;
printf("Sending computation to OpenMP...");
// Pass the raw memory pointers to the C function
step_physics_usm(n, &pos[0], &vel[0], dt, &ranOnDevice);
// Verify the results
if (ranOnDevice == 1)
printf("[SUCCESS] Computation successfully offloaded to the APU!");
else {
printf("[WARNING] Computation fell back to the Host CPU.\n");
printf(" Check your compiler offload flags, OpenMP installation, and GPU drivers.\n");
}
printf("Sample result: pos[0] = %f\n", pos[0]);
return 0;
}