[Openmp-commits] [llvm] [openmp] [OpenMP][Runtime] Handling crash with `OMP_TARGET_OFFLOAD=DISABLED` and invoking `omp_get_default_device()` (PR #171789)
Amit Tiwari via Openmp-commits
openmp-commits at lists.llvm.org
Thu Dec 18 07:33:11 PST 2025
https://github.com/amitamd7 updated https://github.com/llvm/llvm-project/pull/171789
>From ead6e7e8221bc82d5e470a2a652ade91f92008f5 Mon Sep 17 00:00:00 2001
From: amtiwari <amtiwari at amd.com>
Date: Thu, 11 Dec 2025 04:39:45 -0500
Subject: [PATCH 1/2] offloading_disabled_resolve
---
.../test/api/omp_get_default_device_test.c | 37 +++++++
.../target_offload_disabled_default_device.c | 45 ++++++++
openmp/runtime/src/kmp_ftn_entry.h | 19 ++--
...p_target_offload_default_device_combined.c | 63 +++++++++++
...omp_target_offload_default_device_nested.c | 103 ++++++++++++++++++
...target_offload_default_device_operations.c | 89 +++++++++++++++
...omp_target_offload_default_device_simple.c | 47 ++++++++
...omp_target_offload_default_device_target.c | 69 ++++++++++++
...mp_target_offload_default_device_threads.c | 89 +++++++++++++++
9 files changed, 553 insertions(+), 8 deletions(-)
create mode 100644 offload/test/api/omp_get_default_device_test.c
create mode 100644 offload/test/offloading/target_offload_disabled_default_device.c
create mode 100644 openmp/runtime/test/env/omp_target_offload_default_device_combined.c
create mode 100644 openmp/runtime/test/env/omp_target_offload_default_device_nested.c
create mode 100644 openmp/runtime/test/env/omp_target_offload_default_device_operations.c
create mode 100644 openmp/runtime/test/env/omp_target_offload_default_device_simple.c
create mode 100644 openmp/runtime/test/env/omp_target_offload_default_device_target.c
create mode 100644 openmp/runtime/test/env/omp_target_offload_default_device_threads.c
diff --git a/offload/test/api/omp_get_default_device_test.c b/offload/test/api/omp_get_default_device_test.c
new file mode 100644
index 0000000000000..b1ca66cd57188
--- /dev/null
+++ b/offload/test/api/omp_get_default_device_test.c
@@ -0,0 +1,37 @@
+// RUN: %libomptarget-compile-generic
+// RUN: env OMP_TARGET_OFFLOAD=disabled %libomptarget-run-generic 2>&1 |
+// %fcheck-generic
+//
+// Test omp_get_default_device() API behavior when offload is disabled
+
+#include <omp.h>
+#include <stdio.h>
+
+int main() {
+ // Test 1: Default behavior
+ int dev1 = omp_get_default_device();
+ // CHECK: Test 1: {{0}}
+ printf("Test 1: %d\n", dev1);
+
+ // Test 2: After setting device
+ omp_set_default_device(3);
+ int dev2 = omp_get_default_device();
+ // CHECK: Test 2: {{0}}
+ printf("Test 2: %d\n", dev2);
+
+ // Test 3: Multiple sets
+ for (int i = 0; i < 5; i++) {
+ omp_set_default_device(i + 10);
+ int dev = omp_get_default_device();
+ // CHECK: Test 3.{{[0-4]}}: {{0}}
+ printf("Test 3.%d: %d\n", i, dev);
+ }
+
+ // Test 4: Consistency with initial device
+ int initial = omp_get_initial_device();
+ int default_dev = omp_get_default_device();
+ // CHECK: Test 4: EQUAL
+ printf("Test 4: %s\n", (initial == default_dev) ? "EQUAL" : "NOT_EQUAL");
+
+ return 0;
+}
diff --git a/offload/test/offloading/target_offload_disabled_default_device.c b/offload/test/offloading/target_offload_disabled_default_device.c
new file mode 100644
index 0000000000000..20b40a43bb700
--- /dev/null
+++ b/offload/test/offloading/target_offload_disabled_default_device.c
@@ -0,0 +1,45 @@
+// RUN: %libomptarget-compile-generic
+// RUN: env OMP_TARGET_OFFLOAD=disabled %libomptarget-run-generic 2>&1 |
+// %fcheck-generic
+//
+// Test that setting default device before disabling offload doesn't crash
+
+#include <omp.h>
+#include <stdio.h>
+
+int main() {
+ // Set high default device number
+ omp_set_default_device(5);
+
+ // This simulates OMP_TARGET_OFFLOAD=disabled being set after device is chosen
+ // In practice, the environment variable is read at runtime init
+
+ // CHECK: num_devices: 0
+ printf("num_devices: %d\n", omp_get_num_devices());
+
+ // CHECK: initial_device: 0
+ printf("initial_device: %d\n", omp_get_initial_device());
+
+ // CHECK: default_device: 0
+ printf("default_device: %d\n", omp_get_default_device());
+
+ // Target region should execute on host
+ int result = -1;
+#pragma omp target map(from : result)
+ {
+ result = omp_get_device_num();
+ }
+
+ // CHECK: executed_on: 0
+ printf("executed_on: %d\n", result);
+
+ // CHECK: PASS
+ if (result == omp_get_initial_device() &&
+ omp_get_default_device() == omp_get_initial_device()) {
+ printf("PASS\n");
+ return 0;
+ }
+
+ printf("FAIL\n");
+ return 1;
+}
diff --git a/openmp/runtime/src/kmp_ftn_entry.h b/openmp/runtime/src/kmp_ftn_entry.h
index 6adf60e7ad210..4e7c2822b9358 100644
--- a/openmp/runtime/src/kmp_ftn_entry.h
+++ b/openmp/runtime/src/kmp_ftn_entry.h
@@ -1143,14 +1143,6 @@ int FTN_STDCALL KMP_EXPAND_NAME(FTN_GET_TEAM_NUM)(void) {
#endif
}
-int FTN_STDCALL KMP_EXPAND_NAME(FTN_GET_DEFAULT_DEVICE)(void) {
-#if KMP_MIC || KMP_OS_DARWIN || defined(KMP_STUB)
- return 0;
-#else
- return __kmp_entry_thread()->th.th_current_task->td_icvs.default_device;
-#endif
-}
-
void FTN_STDCALL KMP_EXPAND_NAME(FTN_SET_DEFAULT_DEVICE)(int KMP_DEREF arg) {
#if KMP_MIC || KMP_OS_DARWIN || defined(KMP_STUB)
// Nothing.
@@ -1197,6 +1189,17 @@ int FTN_STDCALL KMP_EXPAND_NAME(FTN_GET_INITIAL_DEVICE)(void) {
return KMP_EXPAND_NAME(FTN_GET_NUM_DEVICES)();
}
+int FTN_STDCALL KMP_EXPAND_NAME(FTN_GET_DEFAULT_DEVICE)(void) {
+#if KMP_MIC || KMP_OS_DARWIN || defined(KMP_STUB)
+ return 0;
+#else
+ // When offloading is disabled, return the initial device (host)
+ if (__kmp_target_offload == tgt_disabled)
+ return KMP_EXPAND_NAME(FTN_GET_INITIAL_DEVICE)();
+ return __kmp_entry_thread()->th.th_current_task->td_icvs.default_device;
+#endif
+}
+
#if defined(KMP_STUB)
// Entries for stubs library
// As all *target* functions are C-only parameters always passed by value
diff --git a/openmp/runtime/test/env/omp_target_offload_default_device_combined.c b/openmp/runtime/test/env/omp_target_offload_default_device_combined.c
new file mode 100644
index 0000000000000..4bd8c57bc5785
--- /dev/null
+++ b/openmp/runtime/test/env/omp_target_offload_default_device_combined.c
@@ -0,0 +1,63 @@
+// RUN: %libomp-compile-and-run
+// REQUIRES: ompt
+//
+// Test that omp_get_default_device() returns the initial device (0) when
+// OMP_TARGET_OFFLOAD=DISABLED, with both OMP_DEFAULT_DEVICE environment
+// variable and omp_set_default_device() API call setting non-zero values.
+
+#include <stdio.h>
+#include <stdlib.h>
+#include <omp.h>
+
+extern void kmp_set_defaults(char const *str);
+
+int main() {
+ // Simulate worst case: both environment variable and API call set non-zero
+ // device
+ kmp_set_defaults("OMP_DEFAULT_DEVICE=3");
+ kmp_set_defaults("OMP_TARGET_OFFLOAD=DISABLED");
+
+// Force parallel region to initialize runtime
+#pragma omp parallel
+ {
+ }
+
+ int initial_device = omp_get_initial_device();
+ int default_device_1 = omp_get_default_device();
+
+ printf("With OMP_DEFAULT_DEVICE=3 and OMP_TARGET_OFFLOAD=DISABLED:\n");
+ printf(" initial_device = %d\n", initial_device);
+ printf(" default_device = %d\n", default_device_1);
+
+ if (default_device_1 != initial_device) {
+ fprintf(stderr,
+ "FAIL: Environment variable not overridden by offload disabled\n");
+ return EXIT_FAILURE;
+ }
+
+ // Now also call omp_set_default_device()
+ omp_set_default_device(7);
+ int default_device_2 = omp_get_default_device();
+
+ printf("After additional omp_set_default_device(7):\n");
+ printf(" default_device = %d\n", default_device_2);
+
+ if (default_device_2 != initial_device) {
+ fprintf(stderr, "FAIL: API call not overridden by offload disabled\n");
+ return EXIT_FAILURE;
+ }
+
+ // Verify consistency across multiple calls
+ for (int i = 0; i < 5; i++) {
+ int dev = omp_get_default_device();
+ if (dev != initial_device) {
+ fprintf(stderr,
+ "FAIL: Inconsistent result on call %d: got %d, expected %d\n", i,
+ dev, initial_device);
+ return EXIT_FAILURE;
+ }
+ }
+
+ printf("PASS: default_device consistently returns initial_device\n");
+ return EXIT_SUCCESS;
+}
diff --git a/openmp/runtime/test/env/omp_target_offload_default_device_nested.c b/openmp/runtime/test/env/omp_target_offload_default_device_nested.c
new file mode 100644
index 0000000000000..6ad120e0bff80
--- /dev/null
+++ b/openmp/runtime/test/env/omp_target_offload_default_device_nested.c
@@ -0,0 +1,103 @@
+// RUN: %libomp-compile-and-run
+// REQUIRES: ompt
+//
+// Test that omp_get_default_device() returns the initial device consistently
+// across nested parallel regions and with ICV inheritance when
+// OMP_TARGET_OFFLOAD=DISABLED.
+
+#include <stdio.h>
+#include <stdlib.h>
+#include <omp.h>
+
+extern void kmp_set_defaults(char const *str);
+
+int check_default_device(const char *context, int expected_initial) {
+ int default_dev = omp_get_default_device();
+ int initial_dev = omp_get_initial_device();
+
+ if (default_dev != initial_dev) {
+ fprintf(stderr, "FAIL [%s]: default=%d, initial=%d\n", context, default_dev,
+ initial_dev);
+ return 1;
+ }
+
+ if (initial_dev != expected_initial) {
+ fprintf(stderr, "FAIL [%s]: initial=%d, expected=%d\n", context,
+ initial_dev, expected_initial);
+ return 1;
+ }
+
+ return 0;
+}
+
+int main() {
+ int errors = 0;
+
+ // Set configuration
+ kmp_set_defaults("OMP_DEFAULT_DEVICE=8");
+ kmp_set_defaults("OMP_TARGET_OFFLOAD=DISABLED");
+
+// Initialize runtime
+#pragma omp parallel
+ {
+ }
+
+ int initial_device = omp_get_initial_device();
+ printf("initial_device = %d\n", initial_device);
+
+ // Test 1: Sequential region
+ errors += check_default_device("sequential", initial_device);
+
+// Test 2: Parallel region
+#pragma omp parallel reduction(+ : errors)
+ {
+ errors += check_default_device("parallel", initial_device);
+
+// Test 3: Nested parallel (if supported)
+#pragma omp parallel reduction(+ : errors) if (omp_get_max_threads() > 2)
+ {
+ errors += check_default_device("nested parallel", initial_device);
+ }
+ }
+
+// Test 4: After modifying in one thread
+#pragma omp parallel num_threads(4) reduction(+ : errors)
+ {
+ int tid = omp_get_thread_num();
+
+ // Each thread tries to set different default device
+ omp_set_default_device(tid + 20);
+
+ // But should still get initial device
+ errors += check_default_device("after thread-local set", initial_device);
+
+#pragma omp barrier
+
+ // Check again after barrier
+ errors += check_default_device("after barrier", initial_device);
+ }
+
+ // Test 5: Back in sequential after all the parallel regions
+ errors += check_default_device("sequential final", initial_device);
+
+ // Test 6: Target region context
+ int target_errors = 0;
+#pragma omp target map(tofrom : target_errors)
+ {
+ int default_dev = omp_get_default_device();
+ int initial_dev = omp_get_initial_device();
+ if (default_dev != initial_dev) {
+ target_errors = 1;
+ }
+ }
+ errors += target_errors;
+
+ if (errors > 0) {
+ fprintf(stderr, "FAIL: %d errors detected\n", errors);
+ return EXIT_FAILURE;
+ }
+
+ printf("PASS: default_device consistently returns initial_device across all "
+ "contexts\n");
+ return EXIT_SUCCESS;
+}
diff --git a/openmp/runtime/test/env/omp_target_offload_default_device_operations.c b/openmp/runtime/test/env/omp_target_offload_default_device_operations.c
new file mode 100644
index 0000000000000..6799b74a67349
--- /dev/null
+++ b/openmp/runtime/test/env/omp_target_offload_default_device_operations.c
@@ -0,0 +1,89 @@
+// RUN: %libomp-compile-and-run
+// REQUIRES: ompt
+//
+// Test that device operations using omp_get_default_device() don't crash
+// when OMP_TARGET_OFFLOAD=DISABLED. This simulates real-world usage where
+// the default device is used for device-specific operations.
+
+#include <stdio.h>
+#include <stdlib.h>
+#include <omp.h>
+
+extern void kmp_set_defaults(char const *str);
+
+int main() {
+ // Simulate the problematic scenario: high default device number + disabled
+ // offload Use API call instead of env var to ensure ICV is set
+ omp_set_default_device(10);
+
+ // Now disable offload
+ kmp_set_defaults("OMP_TARGET_OFFLOAD=DISABLED");
+
+// Force parallel region to initialize runtime
+#pragma omp parallel
+ {
+ }
+
+ int device = omp_get_default_device();
+ int num_devices = omp_get_num_devices();
+ int initial_device = omp_get_initial_device();
+
+ printf("Configuration:\n");
+ printf(" num_devices = %d\n", num_devices);
+ printf(" initial_device = %d\n", initial_device);
+ printf(" default_device = %d\n", device);
+
+ // Verify device is in valid range
+ if (device < 0 || device > num_devices) {
+ fprintf(stderr, "FAIL: default_device (%d) is out of valid range [0, %d]\n",
+ device, num_devices);
+ return EXIT_FAILURE;
+ }
+
+ // Test 1: Check if we're on initial device (should be true when offload
+ // disabled)
+ int is_initial = (device == initial_device);
+ if (!is_initial) {
+ fprintf(stderr,
+ "FAIL: default_device (%d) is not the initial_device (%d)\n",
+ device, initial_device);
+ return EXIT_FAILURE;
+ }
+
+ // Test 2: Use device in target region with device clause
+ // This should not crash even though OMP_DEFAULT_DEVICE=10
+ int result = -1;
+#pragma omp target device(device) map(from : result)
+ {
+ result = omp_get_device_num();
+ }
+
+ printf("Target region executed on device: %d\n", result);
+
+ // When offload is disabled, target should execute on host (initial device)
+ if (result != initial_device) {
+ fprintf(stderr, "FAIL: Target executed on device %d, expected %d\n", result,
+ initial_device);
+ return EXIT_FAILURE;
+ }
+
+ // Test 3: Query device properties using the default device
+ int is_host = (device == initial_device);
+ printf("Device %d is_host: %d\n", device, is_host);
+
+ // Test 4: Verify target region with default device specification
+ int test_value = 0;
+#pragma omp target device(omp_get_default_device()) map(tofrom : test_value)
+ {
+ test_value = 42;
+ }
+
+ if (test_value != 42) {
+ fprintf(stderr,
+ "FAIL: Target region with default device did not execute\n");
+ return EXIT_FAILURE;
+ }
+
+ printf("PASS: All device operations completed without crash\n");
+ return EXIT_SUCCESS;
+}
diff --git a/openmp/runtime/test/env/omp_target_offload_default_device_simple.c b/openmp/runtime/test/env/omp_target_offload_default_device_simple.c
new file mode 100644
index 0000000000000..a4a5479e154b3
--- /dev/null
+++ b/openmp/runtime/test/env/omp_target_offload_default_device_simple.c
@@ -0,0 +1,47 @@
+// RUN: %libomp-compile-and-run
+// REQUIRES: ompt
+//
+// Simple smoke test to verify omp_get_default_device() returns initial device
+// when OMP_TARGET_OFFLOAD=DISABLED with OMP_DEFAULT_DEVICE=2.
+// This is the C equivalent of the Fortran smoke test.
+
+#include <stdio.h>
+#include <stdlib.h>
+#include <omp.h>
+
+extern void kmp_set_defaults(char const *str);
+
+int main() {
+ // Key to reproducing bug: Set default device BEFORE disabling offload
+ // This ensures the ICV contains a non-zero value
+ omp_set_default_device(2);
+
+ // Now disable offload
+ kmp_set_defaults("OMP_TARGET_OFFLOAD=DISABLED");
+
+// Initialize runtime
+#pragma omp parallel
+ {
+ }
+
+ int num_devices = omp_get_num_devices();
+ int initial_device = omp_get_initial_device();
+ int default_device = omp_get_default_device();
+
+ // Print results
+ printf("number of devices %d\n", num_devices);
+ printf("initial device %d\n", initial_device);
+ printf("default device %d\n", default_device);
+
+ // The key test: default device should equal initial device
+ if (initial_device == default_device) {
+ printf("PASS\n");
+ return EXIT_SUCCESS;
+ } else {
+ fprintf(stderr, "FAIL: default_device (%d) != initial_device (%d)\n",
+ default_device, initial_device);
+ fprintf(stderr, "This would cause: device number '%d' out of range\n",
+ default_device);
+ return EXIT_FAILURE;
+ }
+}
diff --git a/openmp/runtime/test/env/omp_target_offload_default_device_target.c b/openmp/runtime/test/env/omp_target_offload_default_device_target.c
new file mode 100644
index 0000000000000..a821a0547dda5
--- /dev/null
+++ b/openmp/runtime/test/env/omp_target_offload_default_device_target.c
@@ -0,0 +1,69 @@
+// RUN: %libomp-compile-and-run
+// REQUIRES: ompt
+//
+// Test that omp_get_default_device() returns the initial device (0) when
+// called from within a target region when OMP_TARGET_OFFLOAD=DISABLED.
+// The target region should execute on the host.
+
+#include <stdio.h>
+#include <stdlib.h>
+#include <omp.h>
+
+extern void kmp_set_defaults(char const *str);
+
+int main() {
+ // Set non-zero default device using API (more direct than env var)
+ omp_set_default_device(4);
+
+ // Now disable offload
+ kmp_set_defaults("OMP_TARGET_OFFLOAD=DISABLED");
+
+// Force parallel region to initialize runtime
+#pragma omp parallel
+ {
+ }
+
+ int initial_device = omp_get_initial_device();
+ int host_default_device = omp_get_default_device();
+ int target_default_device = -1;
+ int target_is_initial = -1;
+
+ printf("Host context:\n");
+ printf(" initial_device = %d\n", initial_device);
+ printf(" default_device = %d\n", host_default_device);
+
+// Call omp_get_default_device() from within target region
+// When offload is disabled, this should execute on host
+#pragma omp target map(from : target_default_device, target_is_initial)
+ {
+ target_default_device = omp_get_default_device();
+ target_is_initial = omp_is_initial_device();
+ }
+
+ printf("Target context (executed on host when offload disabled):\n");
+ printf(" default_device = %d\n", target_default_device);
+ printf(" is_initial_device = %d\n", target_is_initial);
+
+ // When offload is disabled, target region executes on host
+ if (target_is_initial != 1) {
+ fprintf(stderr, "FAIL: Target region did not execute on initial device\n");
+ return EXIT_FAILURE;
+ }
+
+ // Both host and target context should return same device
+ if (host_default_device != initial_device) {
+ fprintf(stderr, "FAIL: Host default_device (%d) != initial_device (%d)\n",
+ host_default_device, initial_device);
+ return EXIT_FAILURE;
+ }
+
+ if (target_default_device != initial_device) {
+ fprintf(stderr, "FAIL: Target default_device (%d) != initial_device (%d)\n",
+ target_default_device, initial_device);
+ return EXIT_FAILURE;
+ }
+
+ printf("PASS: default_device returns initial_device in both host and target "
+ "contexts\n");
+ return EXIT_SUCCESS;
+}
diff --git a/openmp/runtime/test/env/omp_target_offload_default_device_threads.c b/openmp/runtime/test/env/omp_target_offload_default_device_threads.c
new file mode 100644
index 0000000000000..1640f6366f86c
--- /dev/null
+++ b/openmp/runtime/test/env/omp_target_offload_default_device_threads.c
@@ -0,0 +1,89 @@
+// RUN: %libomp-compile-and-run
+// REQUIRES: ompt
+//
+// Test that omp_get_default_device() consistently returns the initial device
+// across multiple threads when OMP_TARGET_OFFLOAD=DISABLED.
+// This ensures thread-safety.
+
+#include <stdio.h>
+#include <stdlib.h>
+#include <omp.h>
+
+extern void kmp_set_defaults(char const *str);
+
+int main() {
+ const int NUM_THREADS = 8;
+ const int NUM_ITERATIONS = 100;
+
+ // Set non-zero default device and disable offload
+ kmp_set_defaults("OMP_DEFAULT_DEVICE=6");
+ kmp_set_defaults("OMP_TARGET_OFFLOAD=DISABLED");
+
+// Force parallel region to initialize runtime
+#pragma omp parallel
+ {
+ }
+
+ int initial_device = omp_get_initial_device();
+ int errors = 0;
+
+ printf("Testing with %d threads, %d iterations each\n", NUM_THREADS,
+ NUM_ITERATIONS);
+ printf("initial_device = %d\n", initial_device);
+
+// Test across multiple parallel regions and threads
+#pragma omp parallel num_threads(NUM_THREADS) reduction(+ : errors)
+ {
+ int tid = omp_get_thread_num();
+
+ for (int i = 0; i < NUM_ITERATIONS; i++) {
+ int default_device = omp_get_default_device();
+
+ if (default_device != initial_device) {
+#pragma omp critical
+ {
+ fprintf(
+ stderr,
+ "FAIL: Thread %d iteration %d: default_device=%d, expected=%d\n",
+ tid, i, default_device, initial_device);
+ }
+ errors++;
+ }
+
+ // Also test after setting default device in each thread
+ if (i % 10 == 0) {
+ omp_set_default_device(tid + 10);
+ default_device = omp_get_default_device();
+
+ if (default_device != initial_device) {
+#pragma omp critical
+ {
+ fprintf(
+ stderr,
+ "FAIL: Thread %d after set: default_device=%d, expected=%d\n",
+ tid, default_device, initial_device);
+ }
+ errors++;
+ }
+ }
+ }
+ }
+
+ if (errors > 0) {
+ fprintf(stderr, "FAIL: %d errors detected across all threads\n", errors);
+ return EXIT_FAILURE;
+ }
+
+ // Final verification
+ int final_device = omp_get_default_device();
+ if (final_device != initial_device) {
+ fprintf(stderr,
+ "FAIL: Final check failed: default_device=%d, expected=%d\n",
+ final_device, initial_device);
+ return EXIT_FAILURE;
+ }
+
+ printf("PASS: default_device consistently returns initial_device across all "
+ "threads\n");
+ return EXIT_SUCCESS;
+}
>From 5ce05f8b0aa235e03359fec3d3547ec9eed6b59b Mon Sep 17 00:00:00 2001
From: amtiwari <amtiwari at amd.com>
Date: Thu, 18 Dec 2025 10:26:46 -0500
Subject: [PATCH 2/2] comments_changes
---
.../test/api/omp_get_default_device_test.c | 44 +++++++++-------
.../target_offload_disabled_default_device.c | 52 +++++++++++--------
...p_target_offload_default_device_combined.c | 1 -
...omp_target_offload_default_device_nested.c | 21 +++-----
...target_offload_default_device_operations.c | 14 ++---
...omp_target_offload_default_device_simple.c | 9 ++--
...omp_target_offload_default_device_target.c | 47 ++++++-----------
...mp_target_offload_default_device_threads.c | 1 -
8 files changed, 86 insertions(+), 103 deletions(-)
diff --git a/offload/test/api/omp_get_default_device_test.c b/offload/test/api/omp_get_default_device_test.c
index b1ca66cd57188..2f28178f5b6e3 100644
--- a/offload/test/api/omp_get_default_device_test.c
+++ b/offload/test/api/omp_get_default_device_test.c
@@ -2,36 +2,40 @@
// RUN: env OMP_TARGET_OFFLOAD=disabled %libomptarget-run-generic 2>&1 |
// %fcheck-generic
//
-// Test omp_get_default_device() API behavior when offload is disabled
+// API contract test: omp_get_default_device() behavior when offload is disabled
#include <omp.h>
#include <stdio.h>
int main() {
- // Test 1: Default behavior
+ // Force runtime initialization to parse environment variables
+#pragma omp parallel
+ {
+ }
+
+ int initial = omp_get_initial_device();
+
+ // Test 1: Default device should initially equal initial device
int dev1 = omp_get_default_device();
- // CHECK: Test 1: {{0}}
- printf("Test 1: %d\n", dev1);
+ // CHECK: Test 1: EQUAL
+ printf("Test 1: %s\n", (dev1 == initial) ? "EQUAL" : "NOT_EQUAL");
- // Test 2: After setting device
+ // Test 2: After setting to 3, get should still return initial device (not 3)
omp_set_default_device(3);
int dev2 = omp_get_default_device();
- // CHECK: Test 2: {{0}}
- printf("Test 2: %d\n", dev2);
-
- // Test 3: Multiple sets
- for (int i = 0; i < 5; i++) {
- omp_set_default_device(i + 10);
- int dev = omp_get_default_device();
- // CHECK: Test 3.{{[0-4]}}: {{0}}
- printf("Test 3.%d: %d\n", i, dev);
- }
+ // CHECK: Test 2: EQUAL
+ printf("Test 2: %s\n", (dev2 == initial) ? "EQUAL" : "NOT_EQUAL");
- // Test 4: Consistency with initial device
- int initial = omp_get_initial_device();
- int default_dev = omp_get_default_device();
- // CHECK: Test 4: EQUAL
- printf("Test 4: %s\n", (initial == default_dev) ? "EQUAL" : "NOT_EQUAL");
+ // Test 3: After setting to 10, get should still return initial device
+ omp_set_default_device(10);
+ int dev3 = omp_get_default_device();
+ // CHECK: Test 3: EQUAL
+ printf("Test 3: %s\n", (dev3 == initial) ? "EQUAL" : "NOT_EQUAL");
+
+ // Test 4: All calls return consistent value
+ // CHECK: Test 4: CONSISTENT
+ printf("Test 4: %s\n",
+ (dev1 == dev2 && dev2 == dev3) ? "CONSISTENT" : "INCONSISTENT");
return 0;
}
diff --git a/offload/test/offloading/target_offload_disabled_default_device.c b/offload/test/offloading/target_offload_disabled_default_device.c
index 20b40a43bb700..2febe63068f98 100644
--- a/offload/test/offloading/target_offload_disabled_default_device.c
+++ b/offload/test/offloading/target_offload_disabled_default_device.c
@@ -2,44 +2,54 @@
// RUN: env OMP_TARGET_OFFLOAD=disabled %libomptarget-run-generic 2>&1 |
// %fcheck-generic
//
-// Test that setting default device before disabling offload doesn't crash
+// Integration test: target region execution when offload is disabled
+// with default device set to invalid value
#include <omp.h>
#include <stdio.h>
int main() {
- // Set high default device number
+ // Force runtime initialization to parse environment variables
+#pragma omp parallel
+ {
+ }
+
+ // Set high default device number that would be invalid
omp_set_default_device(5);
- // This simulates OMP_TARGET_OFFLOAD=disabled being set after device is chosen
- // In practice, the environment variable is read at runtime init
+ int num_devices = omp_get_num_devices();
+ int initial_device = omp_get_initial_device();
+ int default_device = omp_get_default_device();
// CHECK: num_devices: 0
- printf("num_devices: %d\n", omp_get_num_devices());
+ printf("num_devices: %d\n", num_devices);
- // CHECK: initial_device: 0
- printf("initial_device: %d\n", omp_get_initial_device());
+ printf("initial_device: %d\n", initial_device);
+ printf("default_device: %d\n", default_device);
- // CHECK: default_device: 0
- printf("default_device: %d\n", omp_get_default_device());
+ // The key test: default device must equal initial device when offload
+ // disabled CHECK: PASS
+ if (default_device == initial_device) {
+ printf("PASS\n");
+ } else {
+ printf("FAIL: default_device=%d, initial_device=%d\n", default_device,
+ initial_device);
+ return 1;
+ }
- // Target region should execute on host
- int result = -1;
-#pragma omp target map(from : result)
+ // Verify target region executes without crashing when offload is disabled
+ int executed = 0;
+#pragma omp target map(tofrom : executed)
{
- result = omp_get_device_num();
+ executed = 1;
}
- // CHECK: executed_on: 0
- printf("executed_on: %d\n", result);
-
- // CHECK: PASS
- if (result == omp_get_initial_device() &&
- omp_get_default_device() == omp_get_initial_device()) {
- printf("PASS\n");
+ // CHECK: Target executed
+ if (executed) {
+ printf("Target executed\n");
return 0;
}
- printf("FAIL\n");
+ printf("FAIL: Target region did not execute\n");
return 1;
}
diff --git a/openmp/runtime/test/env/omp_target_offload_default_device_combined.c b/openmp/runtime/test/env/omp_target_offload_default_device_combined.c
index 4bd8c57bc5785..fcb4a0eddfb39 100644
--- a/openmp/runtime/test/env/omp_target_offload_default_device_combined.c
+++ b/openmp/runtime/test/env/omp_target_offload_default_device_combined.c
@@ -1,5 +1,4 @@
// RUN: %libomp-compile-and-run
-// REQUIRES: ompt
//
// Test that omp_get_default_device() returns the initial device (0) when
// OMP_TARGET_OFFLOAD=DISABLED, with both OMP_DEFAULT_DEVICE environment
diff --git a/openmp/runtime/test/env/omp_target_offload_default_device_nested.c b/openmp/runtime/test/env/omp_target_offload_default_device_nested.c
index 6ad120e0bff80..ea95ef775ff05 100644
--- a/openmp/runtime/test/env/omp_target_offload_default_device_nested.c
+++ b/openmp/runtime/test/env/omp_target_offload_default_device_nested.c
@@ -1,5 +1,4 @@
// RUN: %libomp-compile-and-run
-// REQUIRES: ompt
//
// Test that omp_get_default_device() returns the initial device consistently
// across nested parallel regions and with ICV inheritance when
@@ -11,7 +10,7 @@
extern void kmp_set_defaults(char const *str);
-int check_default_device(const char *context, int expected_initial) {
+int check_default_device(const char *context) {
int default_dev = omp_get_default_device();
int initial_dev = omp_get_initial_device();
@@ -21,12 +20,6 @@ int check_default_device(const char *context, int expected_initial) {
return 1;
}
- if (initial_dev != expected_initial) {
- fprintf(stderr, "FAIL [%s]: initial=%d, expected=%d\n", context,
- initial_dev, expected_initial);
- return 1;
- }
-
return 0;
}
@@ -46,17 +39,17 @@ int main() {
printf("initial_device = %d\n", initial_device);
// Test 1: Sequential region
- errors += check_default_device("sequential", initial_device);
+ errors += check_default_device("sequential");
// Test 2: Parallel region
#pragma omp parallel reduction(+ : errors)
{
- errors += check_default_device("parallel", initial_device);
+ errors += check_default_device("parallel");
// Test 3: Nested parallel (if supported)
#pragma omp parallel reduction(+ : errors) if (omp_get_max_threads() > 2)
{
- errors += check_default_device("nested parallel", initial_device);
+ errors += check_default_device("nested parallel");
}
}
@@ -69,16 +62,16 @@ int main() {
omp_set_default_device(tid + 20);
// But should still get initial device
- errors += check_default_device("after thread-local set", initial_device);
+ errors += check_default_device("after thread-local set");
#pragma omp barrier
// Check again after barrier
- errors += check_default_device("after barrier", initial_device);
+ errors += check_default_device("after barrier");
}
// Test 5: Back in sequential after all the parallel regions
- errors += check_default_device("sequential final", initial_device);
+ errors += check_default_device("sequential final");
// Test 6: Target region context
int target_errors = 0;
diff --git a/openmp/runtime/test/env/omp_target_offload_default_device_operations.c b/openmp/runtime/test/env/omp_target_offload_default_device_operations.c
index 6799b74a67349..965fb4912938e 100644
--- a/openmp/runtime/test/env/omp_target_offload_default_device_operations.c
+++ b/openmp/runtime/test/env/omp_target_offload_default_device_operations.c
@@ -1,5 +1,4 @@
// RUN: %libomp-compile-and-run
-// REQUIRES: ompt
//
// Test that device operations using omp_get_default_device() don't crash
// when OMP_TARGET_OFFLOAD=DISABLED. This simulates real-world usage where
@@ -12,13 +11,13 @@
extern void kmp_set_defaults(char const *str);
int main() {
+ // Disable offload first to avoid early runtime initialization
+ kmp_set_defaults("OMP_TARGET_OFFLOAD=DISABLED");
+
// Simulate the problematic scenario: high default device number + disabled
// offload Use API call instead of env var to ensure ICV is set
omp_set_default_device(10);
- // Now disable offload
- kmp_set_defaults("OMP_TARGET_OFFLOAD=DISABLED");
-
// Force parallel region to initialize runtime
#pragma omp parallel
{
@@ -60,13 +59,6 @@ int main() {
printf("Target region executed on device: %d\n", result);
- // When offload is disabled, target should execute on host (initial device)
- if (result != initial_device) {
- fprintf(stderr, "FAIL: Target executed on device %d, expected %d\n", result,
- initial_device);
- return EXIT_FAILURE;
- }
-
// Test 3: Query device properties using the default device
int is_host = (device == initial_device);
printf("Device %d is_host: %d\n", device, is_host);
diff --git a/openmp/runtime/test/env/omp_target_offload_default_device_simple.c b/openmp/runtime/test/env/omp_target_offload_default_device_simple.c
index a4a5479e154b3..136a97d4440a5 100644
--- a/openmp/runtime/test/env/omp_target_offload_default_device_simple.c
+++ b/openmp/runtime/test/env/omp_target_offload_default_device_simple.c
@@ -1,5 +1,4 @@
// RUN: %libomp-compile-and-run
-// REQUIRES: ompt
//
// Simple smoke test to verify omp_get_default_device() returns initial device
// when OMP_TARGET_OFFLOAD=DISABLED with OMP_DEFAULT_DEVICE=2.
@@ -12,13 +11,13 @@
extern void kmp_set_defaults(char const *str);
int main() {
- // Key to reproducing bug: Set default device BEFORE disabling offload
+ // Disable offload first to avoid early runtime initialization
+ kmp_set_defaults("OMP_TARGET_OFFLOAD=DISABLED");
+
+ // Key to reproducing bug: Set default device to non-zero value
// This ensures the ICV contains a non-zero value
omp_set_default_device(2);
- // Now disable offload
- kmp_set_defaults("OMP_TARGET_OFFLOAD=DISABLED");
-
// Initialize runtime
#pragma omp parallel
{
diff --git a/openmp/runtime/test/env/omp_target_offload_default_device_target.c b/openmp/runtime/test/env/omp_target_offload_default_device_target.c
index a821a0547dda5..44e065eaa1845 100644
--- a/openmp/runtime/test/env/omp_target_offload_default_device_target.c
+++ b/openmp/runtime/test/env/omp_target_offload_default_device_target.c
@@ -1,9 +1,7 @@
// RUN: %libomp-compile-and-run
-// REQUIRES: ompt
//
-// Test that omp_get_default_device() returns the initial device (0) when
-// called from within a target region when OMP_TARGET_OFFLOAD=DISABLED.
-// The target region should execute on the host.
+// Test that omp_get_default_device() returns the initial device when
+// OMP_TARGET_OFFLOAD=DISABLED, and that target regions execute on the host.
#include <stdio.h>
#include <stdlib.h>
@@ -12,12 +10,12 @@
extern void kmp_set_defaults(char const *str);
int main() {
+ // Disable offload first to avoid early runtime initialization
+ kmp_set_defaults("OMP_TARGET_OFFLOAD=DISABLED");
+
// Set non-zero default device using API (more direct than env var)
omp_set_default_device(4);
- // Now disable offload
- kmp_set_defaults("OMP_TARGET_OFFLOAD=DISABLED");
-
// Force parallel region to initialize runtime
#pragma omp parallel
{
@@ -25,45 +23,34 @@ int main() {
int initial_device = omp_get_initial_device();
int host_default_device = omp_get_default_device();
- int target_default_device = -1;
int target_is_initial = -1;
printf("Host context:\n");
printf(" initial_device = %d\n", initial_device);
printf(" default_device = %d\n", host_default_device);
-// Call omp_get_default_device() from within target region
-// When offload is disabled, this should execute on host
-#pragma omp target map(from : target_default_device, target_is_initial)
+ // Verify default_device returns initial_device in host context
+ if (host_default_device != initial_device) {
+ fprintf(stderr, "FAIL: Host default_device (%d) != initial_device (%d)\n",
+ host_default_device, initial_device);
+ return EXIT_FAILURE;
+ }
+
+ // Verify target region executes on host when offload is disabled
+#pragma omp target map(from : target_is_initial)
{
- target_default_device = omp_get_default_device();
target_is_initial = omp_is_initial_device();
}
- printf("Target context (executed on host when offload disabled):\n");
- printf(" default_device = %d\n", target_default_device);
+ printf("Target region:\n");
printf(" is_initial_device = %d\n", target_is_initial);
- // When offload is disabled, target region executes on host
if (target_is_initial != 1) {
fprintf(stderr, "FAIL: Target region did not execute on initial device\n");
return EXIT_FAILURE;
}
- // Both host and target context should return same device
- if (host_default_device != initial_device) {
- fprintf(stderr, "FAIL: Host default_device (%d) != initial_device (%d)\n",
- host_default_device, initial_device);
- return EXIT_FAILURE;
- }
-
- if (target_default_device != initial_device) {
- fprintf(stderr, "FAIL: Target default_device (%d) != initial_device (%d)\n",
- target_default_device, initial_device);
- return EXIT_FAILURE;
- }
-
- printf("PASS: default_device returns initial_device in both host and target "
- "contexts\n");
+ printf("PASS: default_device returns initial_device and target executes on "
+ "host\n");
return EXIT_SUCCESS;
}
diff --git a/openmp/runtime/test/env/omp_target_offload_default_device_threads.c b/openmp/runtime/test/env/omp_target_offload_default_device_threads.c
index 1640f6366f86c..d93b4d0a75f94 100644
--- a/openmp/runtime/test/env/omp_target_offload_default_device_threads.c
+++ b/openmp/runtime/test/env/omp_target_offload_default_device_threads.c
@@ -1,5 +1,4 @@
// RUN: %libomp-compile-and-run
-// REQUIRES: ompt
//
// Test that omp_get_default_device() consistently returns the initial device
// across multiple threads when OMP_TARGET_OFFLOAD=DISABLED.
More information about the Openmp-commits
mailing list