[llvm] [openmp] [OFFLOAD] Add spirv implementation for named barrier (PR #180393)
via llvm-commits
llvm-commits at lists.llvm.org
Thu Mar 26 16:52:15 PDT 2026
https://github.com/fineg74 updated https://github.com/llvm/llvm-project/pull/180393
>From 2ddf1f718c7319c5b18efb9102ca4267b481e10e Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Sat, 7 Feb 2026 23:16:41 -0800
Subject: [PATCH 01/17] Add spirv implementation for named barrier
---
openmp/device/src/Synchronization.cpp | 43 +++++++++++++++++++++++++--
1 file changed, 41 insertions(+), 2 deletions(-)
diff --git a/openmp/device/src/Synchronization.cpp b/openmp/device/src/Synchronization.cpp
index 3554226d2ee75..7a065b0a27fcf 100644
--- a/openmp/device/src/Synchronization.cpp
+++ b/openmp/device/src/Synchronization.cpp
@@ -181,8 +181,47 @@ void setCriticalLock(omp_lock_t *Lock) { setLock(Lock); }
///}
#if defined(__SPIRV__)
-void namedBarrierInit() { __builtin_trap(); } // TODO
-void namedBarrier() { __builtin_trap(); } // TODO
+
+[[clang::loader_uninitialized]] Local<uint32_t> namedBarrierTracker;
+
+void namedBarrierInit() {
+ atomic::store(&namedBarrierTracker, 0u, atomic::seq_cst);
+}
+
+void namedBarrier() {
+ uint32_t NumThreads = omp_get_num_threads();
+
+ // Uses two 16 bit unsigned counters. One for the number of threads to have
+ // reached the barrier, and one to count how many times the barrier has been
+ // passed. These are packed in a single atomically accessed 32 bit integer.
+ // Low bits for the number of threads, assumed zero before this call.
+ // High bits to count the number of times the barrier has been passed.
+
+ // Increment the low 16 bits once.
+
+ uint32_t load = atomic::add(&namedBarrierTracker, 1,
+ atomic::seq_cst);
+
+ // Record the number of times the barrier has been passed
+ uint32_t generation = load & 0xffff0000u;
+
+ if ((load & 0x0000ffffu) == (NumThreads - 1)) {
+ // Reached NumWaves in low bits so this is the last wave.
+ // Set low bits to zero and increment high bits
+ load += 0x00010000u; // wrap is safe
+ load &= 0xffff0000u; // because bits zeroed second
+
+ // Reset the wave counter and release the waiting waves
+ atomic::store(&namedBarrierTracker, load, atomic::seq_cst);
+ } else {
+ // more waves still to go, spin until generation counter changes
+ do {
+ load = atomic::load(&namedBarrierTracker, atomic::seq_cst);
+ } while ((load & 0xffff0000u) == generation);
+ }
+ __gpu_sync_threads();
+
+}
void unsetLock(omp_lock_t *Lock) {
atomic::store((int32_t *)Lock, 0, atomic::seq_cst);
>From 7ada5b0b4a848192b914873639feb54211f314f6 Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Sat, 7 Feb 2026 23:23:01 -0800
Subject: [PATCH 02/17] Fix formatting
---
openmp/device/src/Synchronization.cpp | 4 +---
1 file changed, 1 insertion(+), 3 deletions(-)
diff --git a/openmp/device/src/Synchronization.cpp b/openmp/device/src/Synchronization.cpp
index 7a065b0a27fcf..cebc9ea2e5796 100644
--- a/openmp/device/src/Synchronization.cpp
+++ b/openmp/device/src/Synchronization.cpp
@@ -199,8 +199,7 @@ void namedBarrier() {
// Increment the low 16 bits once.
- uint32_t load = atomic::add(&namedBarrierTracker, 1,
- atomic::seq_cst);
+ uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst);
// Record the number of times the barrier has been passed
uint32_t generation = load & 0xffff0000u;
@@ -220,7 +219,6 @@ void namedBarrier() {
} while ((load & 0xffff0000u) == generation);
}
__gpu_sync_threads();
-
}
void unsetLock(omp_lock_t *Lock) {
>From b138f5115021ee19710531b5209455b2e2014b11 Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Sat, 7 Feb 2026 23:28:46 -0800
Subject: [PATCH 03/17] Fix formatting
---
openmp/device/src/Synchronization.cpp | 4 ++--
1 file changed, 2 insertions(+), 2 deletions(-)
diff --git a/openmp/device/src/Synchronization.cpp b/openmp/device/src/Synchronization.cpp
index cebc9ea2e5796..1053ccf0ea092 100644
--- a/openmp/device/src/Synchronization.cpp
+++ b/openmp/device/src/Synchronization.cpp
@@ -190,7 +190,7 @@ void namedBarrierInit() {
void namedBarrier() {
uint32_t NumThreads = omp_get_num_threads();
-
+
// Uses two 16 bit unsigned counters. One for the number of threads to have
// reached the barrier, and one to count how many times the barrier has been
// passed. These are packed in a single atomically accessed 32 bit integer.
@@ -199,7 +199,7 @@ void namedBarrier() {
// Increment the low 16 bits once.
- uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst);
+ uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst);
// Record the number of times the barrier has been passed
uint32_t generation = load & 0xffff0000u;
>From 23cc317881c3feaf42a7ce6012be19529d79f59c Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Wed, 25 Feb 2026 12:30:22 -0800
Subject: [PATCH 04/17] Simplify implementation
---
openmp/device/src/Synchronization.cpp | 36 ++++++++-------------------
1 file changed, 10 insertions(+), 26 deletions(-)
diff --git a/openmp/device/src/Synchronization.cpp b/openmp/device/src/Synchronization.cpp
index 1053ccf0ea092..8734467f7b360 100644
--- a/openmp/device/src/Synchronization.cpp
+++ b/openmp/device/src/Synchronization.cpp
@@ -190,33 +190,17 @@ void namedBarrierInit() {
void namedBarrier() {
uint32_t NumThreads = omp_get_num_threads();
+ uint32_t ThreadId = mapping::getThreadIdInBlock();
+ if (ThreadId < NumThreads) {
+ uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst);
- // Uses two 16 bit unsigned counters. One for the number of threads to have
- // reached the barrier, and one to count how many times the barrier has been
- // passed. These are packed in a single atomically accessed 32 bit integer.
- // Low bits for the number of threads, assumed zero before this call.
- // High bits to count the number of times the barrier has been passed.
-
- // Increment the low 16 bits once.
-
- uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst);
-
- // Record the number of times the barrier has been passed
- uint32_t generation = load & 0xffff0000u;
-
- if ((load & 0x0000ffffu) == (NumThreads - 1)) {
- // Reached NumWaves in low bits so this is the last wave.
- // Set low bits to zero and increment high bits
- load += 0x00010000u; // wrap is safe
- load &= 0xffff0000u; // because bits zeroed second
-
- // Reset the wave counter and release the waiting waves
- atomic::store(&namedBarrierTracker, load, atomic::seq_cst);
- } else {
- // more waves still to go, spin until generation counter changes
- do {
- load = atomic::load(&namedBarrierTracker, atomic::seq_cst);
- } while ((load & 0xffff0000u) == generation);
+ if (load == NumThreads - 1) {
+ atomic::store(&namedBarrierTracker, 0, atomic::seq_cst);
+ } else {
+ do {
+ load = atomic::load(&namedBarrierTracker, atomic::seq_cst);
+ } while (load != 0);
+ }
}
__gpu_sync_threads();
}
>From fd2a8b4233162d8d93c07e3892daf5aeb2c84f0a Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Wed, 4 Mar 2026 14:16:32 -0800
Subject: [PATCH 05/17] Tweak number of threads calculation
---
openmp/device/src/Synchronization.cpp | 2 +-
1 file changed, 1 insertion(+), 1 deletion(-)
diff --git a/openmp/device/src/Synchronization.cpp b/openmp/device/src/Synchronization.cpp
index 8734467f7b360..dee36b7e56027 100644
--- a/openmp/device/src/Synchronization.cpp
+++ b/openmp/device/src/Synchronization.cpp
@@ -189,7 +189,7 @@ void namedBarrierInit() {
}
void namedBarrier() {
- uint32_t NumThreads = omp_get_num_threads();
+ uint32_t NumThreads = mapping::getMaxTeamThreads();
uint32_t ThreadId = mapping::getThreadIdInBlock();
if (ThreadId < NumThreads) {
uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst);
>From b5ba18bcfc412d9e2271fa0663f657f38d8e442b Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Wed, 4 Mar 2026 15:54:53 -0800
Subject: [PATCH 06/17] Revert "Tweak number of threads calculation"
This reverts commit fd2a8b4233162d8d93c07e3892daf5aeb2c84f0a.
---
openmp/device/src/Synchronization.cpp | 2 +-
1 file changed, 1 insertion(+), 1 deletion(-)
diff --git a/openmp/device/src/Synchronization.cpp b/openmp/device/src/Synchronization.cpp
index dee36b7e56027..8734467f7b360 100644
--- a/openmp/device/src/Synchronization.cpp
+++ b/openmp/device/src/Synchronization.cpp
@@ -189,7 +189,7 @@ void namedBarrierInit() {
}
void namedBarrier() {
- uint32_t NumThreads = mapping::getMaxTeamThreads();
+ uint32_t NumThreads = omp_get_num_threads();
uint32_t ThreadId = mapping::getThreadIdInBlock();
if (ThreadId < NumThreads) {
uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst);
>From 2553bb43399865b1e9b3e8779f0f863701897458 Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Wed, 4 Mar 2026 18:06:33 -0800
Subject: [PATCH 07/17] Revert "Revert "Tweak number of threads calculation""
This reverts commit b5ba18bcfc412d9e2271fa0663f657f38d8e442b.
---
openmp/device/src/Synchronization.cpp | 2 +-
1 file changed, 1 insertion(+), 1 deletion(-)
diff --git a/openmp/device/src/Synchronization.cpp b/openmp/device/src/Synchronization.cpp
index 8734467f7b360..dee36b7e56027 100644
--- a/openmp/device/src/Synchronization.cpp
+++ b/openmp/device/src/Synchronization.cpp
@@ -189,7 +189,7 @@ void namedBarrierInit() {
}
void namedBarrier() {
- uint32_t NumThreads = omp_get_num_threads();
+ uint32_t NumThreads = mapping::getMaxTeamThreads();
uint32_t ThreadId = mapping::getThreadIdInBlock();
if (ThreadId < NumThreads) {
uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst);
>From 98971b72ba669c0c3f8ae628f76e8da8f3750125 Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Fri, 6 Mar 2026 16:42:30 -0800
Subject: [PATCH 08/17] Simplify the implementation
---
openmp/device/src/Synchronization.cpp | 21 +++++++++------------
1 file changed, 9 insertions(+), 12 deletions(-)
diff --git a/openmp/device/src/Synchronization.cpp b/openmp/device/src/Synchronization.cpp
index dee36b7e56027..0fa6abd751f33 100644
--- a/openmp/device/src/Synchronization.cpp
+++ b/openmp/device/src/Synchronization.cpp
@@ -189,18 +189,15 @@ void namedBarrierInit() {
}
void namedBarrier() {
- uint32_t NumThreads = mapping::getMaxTeamThreads();
- uint32_t ThreadId = mapping::getThreadIdInBlock();
- if (ThreadId < NumThreads) {
- uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst);
-
- if (load == NumThreads - 1) {
- atomic::store(&namedBarrierTracker, 0, atomic::seq_cst);
- } else {
- do {
- load = atomic::load(&namedBarrierTracker, atomic::seq_cst);
- } while (load != 0);
- }
+ uint32_t NumThreads = omp_get_num_threads();
+ uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst);
+
+ if (load == NumThreads - 1) {
+ atomic::store(&namedBarrierTracker, 0, atomic::seq_cst);
+ } else {
+ do {
+ load = atomic::load(&namedBarrierTracker, atomic::seq_cst);
+ } while (load != 0);
}
__gpu_sync_threads();
}
>From cc9cee4dbce84bf936895ecfbac185f77546cedd Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Tue, 10 Mar 2026 09:14:41 -0700
Subject: [PATCH 09/17] Disable hanging tests
---
offload/test/mapping/firstprivate_aligned.cpp | 1 +
offload/test/offloading/atomic-compare-signedness.c | 2 +-
offload/test/offloading/bug49021.cpp | 2 +-
offload/test/sanitizer/kernel_trap_async.c | 2 +-
offload/test/sanitizer/use_after_free_1.c | 2 +-
offload/test/sanitizer/use_after_free_2.c | 2 +-
6 files changed, 6 insertions(+), 5 deletions(-)
diff --git a/offload/test/mapping/firstprivate_aligned.cpp b/offload/test/mapping/firstprivate_aligned.cpp
index ae6be0f0c07f4..4521f4397112b 100644
--- a/offload/test/mapping/firstprivate_aligned.cpp
+++ b/offload/test/mapping/firstprivate_aligned.cpp
@@ -1,4 +1,5 @@
// RUN: %libomptarget-compilexx-generic -O3 && %libomptarget-run-generic
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/offloading/atomic-compare-signedness.c b/offload/test/offloading/atomic-compare-signedness.c
index e1c8a2f846a8f..2fd6a484460d0 100644
--- a/offload/test/offloading/atomic-compare-signedness.c
+++ b/offload/test/offloading/atomic-compare-signedness.c
@@ -7,7 +7,7 @@
// RUN: %libomptarget-run-generic | %fcheck-generic
// RUN: %libomptarget-compileopt-generic -fopenmp-version=51
// RUN: %libomptarget-run-generic | %fcheck-generic
-// XFAIL: intelgpu
+// UNSUPPORTED: intelgpu
// High parallelism increases our chances of detecting a lack of atomicity.
#define NUM_THREADS_TRY 256
diff --git a/offload/test/offloading/bug49021.cpp b/offload/test/offloading/bug49021.cpp
index 1bd35c66f5db2..6d919871f0d26 100644
--- a/offload/test/offloading/bug49021.cpp
+++ b/offload/test/offloading/bug49021.cpp
@@ -3,7 +3,7 @@
// RUN: %libomptarget-compilexx-generic -O3 -ffast-math && %libomptarget-run-generic
// RUN: %libomptarget-compileoptxx-generic -O3 && %libomptarget-run-generic
// RUN: %libomptarget-compileoptxx-generic -O3 -ffast-math && %libomptarget-run-generic
-// XFAIL: intelgpu
+// UNSUPPORTED: intelgpu
// clang-format on
#include <iostream>
diff --git a/offload/test/sanitizer/kernel_trap_async.c b/offload/test/sanitizer/kernel_trap_async.c
index ae1bb8d322e4c..aa31c6371fa0c 100644
--- a/offload/test/sanitizer/kernel_trap_async.c
+++ b/offload/test/sanitizer/kernel_trap_async.c
@@ -12,7 +12,7 @@
// UNSUPPORTED: aarch64-unknown-linux-gnu
// UNSUPPORTED: x86_64-unknown-linux-gnu
// UNSUPPORTED: s390x-ibm-linux-gnu
-// XFAIL: intelgpu
+// UNSUPPORTED: intelgpu
#include <omp.h>
diff --git a/offload/test/sanitizer/use_after_free_1.c b/offload/test/sanitizer/use_after_free_1.c
index 927e54ddab2bd..2dee3d107a115 100644
--- a/offload/test/sanitizer/use_after_free_1.c
+++ b/offload/test/sanitizer/use_after_free_1.c
@@ -9,7 +9,7 @@
// UNSUPPORTED: nvidiagpu
//
// REQUIRES: gpu
-// XFAIL: intelgpu
+// UNSUPPORTED: intelgpu
#include <omp.h>
diff --git a/offload/test/sanitizer/use_after_free_2.c b/offload/test/sanitizer/use_after_free_2.c
index ece0cfe60875d..3ba621524e566 100644
--- a/offload/test/sanitizer/use_after_free_2.c
+++ b/offload/test/sanitizer/use_after_free_2.c
@@ -7,7 +7,7 @@
// UNSUPPORTED: nvidiagpu
//
// REQUIRES: gpu
-// XFAIL: intelgpu
+// UNSUPPORTED: intelgpu
// If offload memory pooling is enabled for a large allocation, reuse error is
// not detected. UNSUPPORTED: large_allocation_memory_pool
>From 0edfe972ff4bb69798b2c142ca7e471cf234bbb7 Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Wed, 11 Mar 2026 21:06:55 -0700
Subject: [PATCH 10/17] Disable hanging tests
---
offload/test/api/omp_device_uid.c | 3 ++-
offload/test/api/omp_get_device_num.c | 2 ++
offload/test/api/omp_get_mapped_ptr.c | 3 ++-
offload/test/api/omp_get_num_devices.c | 2 ++
offload/test/api/omp_get_num_procs.c | 3 ++-
offload/test/api/omp_indirect_call.c | 3 ++-
offload/test/api/ompx_3d.c | 2 ++
offload/test/api/ompx_3d.cpp | 2 ++
offload/test/mapping/array_section_implicit_capture.c | 3 ++-
offload/test/mapping/chained_containing_structs_1.cc | 3 ++-
offload/test/mapping/chained_containing_structs_2.cc | 3 ++-
offload/test/mapping/chained_containing_structs_3.cc | 3 ++-
offload/test/mapping/data_member_ref.cpp | 3 ++-
offload/test/mapping/declare_mapper_nested_default_mappers.cpp | 3 ++-
.../test/mapping/declare_mapper_nested_default_mappers_1.cpp | 3 ++-
offload/test/mapping/declare_mapper_nested_mappers.cpp | 3 ++-
offload/test/mapping/declare_mapper_target_checks.cpp | 3 ++-
offload/test/mapping/delete_inf_refcount.c | 2 ++
offload/test/mapping/duplicate_mappings_1.cpp | 3 ++-
offload/test/mapping/duplicate_mappings_2.cpp | 3 ++-
offload/test/mapping/firstprivate_aligned.cpp | 1 +
offload/test/mapping/implicit_device_ptr.c | 3 ++-
offload/test/mapping/low_alignment.c | 3 ++-
offload/test/mapping/map_both_pointer_pointee.c | 3 ++-
offload/test/mapping/map_ordering_tgt_alloc_tofrom.c | 3 ++-
offload/test/mapping/map_ordering_tgt_data_alloc_from.c | 3 ++-
offload/test/mapping/map_ordering_tgt_data_alloc_to_from.c | 3 ++-
offload/test/mapping/map_ordering_tgt_data_alloc_tofrom.c | 3 ++-
offload/test/mapping/map_ordering_tgt_exit_data_delete_from.c | 3 ++-
offload/test/mapping/ompx_hold/struct.c | 3 ++-
offload/test/mapping/ompx_hold/target.c | 3 ++-
offload/test/mapping/pr38704.c | 3 ++-
offload/test/mapping/present/target.c | 2 ++
offload/test/mapping/present/target_array_extension.c | 2 ++
offload/test/mapping/present/unified_shared_memory.c | 2 ++
offload/test/mapping/present/zero_length_array_section.c | 2 ++
offload/test/mapping/present/zero_length_array_section_exit.c | 2 ++
offload/test/mapping/ptr_and_obj_motion.c | 3 ++-
offload/test/mapping/target_data_array_extension_at_exit.c | 3 ++-
offload/test/mapping/target_derefence_array_pointrs.cpp | 3 ++-
offload/test/mapping/target_has_device_addr.c | 3 ++-
offload/test/mapping/target_implicit_partial_map.c | 3 ++-
offload/test/mapping/target_map_for_member_data.cpp | 3 ++-
offload/test/offloading/atomic-compare-signedness.c | 1 +
offload/test/offloading/bug47654.cpp | 3 ++-
offload/test/offloading/bug49021.cpp | 1 +
offload/test/offloading/bug50022.cpp | 3 ++-
offload/test/offloading/bug53727.cpp | 3 ++-
offload/test/offloading/bug64959.c | 2 ++
offload/test/offloading/ctor_dtor.cpp | 3 ++-
offload/test/offloading/dynamic_module_load.c | 2 ++
offload/test/offloading/extern.c | 3 ++-
offload/test/offloading/force-usm.cpp | 3 ++-
offload/test/offloading/global_constructor.cpp | 3 ++-
offload/test/offloading/host_as_target.c | 3 ++-
offload/test/offloading/info.c | 2 ++
offload/test/offloading/memory_manager.cpp | 2 ++
offload/test/offloading/offloading_success.c | 2 ++
offload/test/offloading/offloading_success.cpp | 2 ++
offload/test/offloading/parallel_offloading_map.cpp | 2 ++
offload/test/offloading/requires.c | 2 ++
offload/test/offloading/static_linking.c | 3 ++-
offload/test/offloading/strided_multiple_update_from.c | 3 ++-
offload/test/offloading/strided_multiple_update_to.c | 3 ++-
offload/test/offloading/strided_partial_update_from.c | 3 ++-
offload/test/offloading/strided_partial_update_to.c | 3 ++-
offload/test/offloading/strided_ptr_multiple_update_from.c | 3 ++-
offload/test/offloading/strided_ptr_multiple_update_to.c | 3 ++-
offload/test/offloading/strided_ptr_partial_update_from.c | 3 ++-
offload/test/offloading/strided_ptr_partial_update_to.c | 3 ++-
offload/test/offloading/strided_update_count_expression.c | 2 ++
.../test/offloading/strided_update_count_expression_complex.c | 3 ++-
offload/test/offloading/strided_update_from.c | 3 ++-
.../strided_update_multiple_arrays_count_expression.c | 2 ++
.../strided_update_multiple_arrays_variable_stride.c | 3 ++-
offload/test/offloading/strided_update_to.c | 3 ++-
.../test/offloading/strided_update_variable_count_and_stride.c | 3 ++-
offload/test/offloading/strided_update_variable_stride.c | 2 ++
offload/test/offloading/strided_update_variable_stride_misc.c | 2 ++
offload/test/offloading/target-teams-atomic.c | 2 ++
offload/test/offloading/target-tile.c | 3 ++-
offload/test/offloading/target_depend_nowait.cpp | 3 ++-
offload/test/offloading/target_nowait_target.cpp | 3 ++-
offload/test/offloading/target_update_from.c | 3 ++-
offload/test/offloading/target_update_ptr_count_expression.c | 2 ++
.../offloading/target_update_ptr_variable_count_and_stride.c | 2 ++
offload/test/offloading/target_update_ptr_variable_stride.c | 2 ++
.../offloading/target_update_strided_struct_count_expression.c | 2 ++
offload/test/offloading/target_update_strided_struct_from.c | 3 ++-
.../offloading/target_update_strided_struct_multiple_from.c | 3 ++-
.../test/offloading/target_update_strided_struct_multiple_to.c | 3 ++-
.../offloading/target_update_strided_struct_partial_from.c | 3 ++-
.../test/offloading/target_update_strided_struct_partial_to.c | 3 ++-
offload/test/offloading/target_update_strided_struct_to.c | 3 ++-
.../target_update_strided_struct_variable_count_and_stride.c | 2 ++
.../offloading/target_update_strided_struct_variable_stride.c | 2 ++
offload/test/offloading/target_update_to.c | 3 ++-
offload/test/offloading/task_in_reduction_target.c | 2 ++
offload/test/offloading/workshare_chunk.c | 3 ++-
offload/test/sanitizer/kernel_crash.c | 3 ++-
offload/test/sanitizer/kernel_crash_async.c | 3 ++-
offload/test/sanitizer/kernel_crash_many.c | 3 ++-
offload/test/sanitizer/kernel_trap.c | 3 ++-
offload/test/sanitizer/kernel_trap.cpp | 3 ++-
offload/test/sanitizer/kernel_trap_async.c | 1 +
offload/test/sanitizer/kernel_trap_many.c | 3 ++-
offload/test/sanitizer/use_after_free_1.c | 1 +
offload/test/sanitizer/use_after_free_2.c | 1 +
offload/test/unified_shared_memory/associate_ptr.c | 3 ++-
offload/test/unified_shared_memory/close_manual.c | 2 ++
offload/test/unified_shared_memory/close_member.c | 3 ++-
111 files changed, 216 insertions(+), 74 deletions(-)
diff --git a/offload/test/api/omp_device_uid.c b/offload/test/api/omp_device_uid.c
index 2a41d8d04ef8a..0db41f76a73b2 100644
--- a/offload/test/api/omp_device_uid.c
+++ b/offload/test/api/omp_device_uid.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
#include <string.h>
diff --git a/offload/test/api/omp_get_device_num.c b/offload/test/api/omp_get_device_num.c
index 31e13b0050d41..b5c26cdf0d584 100644
--- a/offload/test/api/omp_get_device_num.c
+++ b/offload/test/api/omp_get_device_num.c
@@ -1,4 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/api/omp_get_mapped_ptr.c b/offload/test/api/omp_get_mapped_ptr.c
index 4436c02259b68..9ba8fc3309edc 100644
--- a/offload/test/api/omp_get_mapped_ptr.c
+++ b/offload/test/api/omp_get_mapped_ptr.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-and-run-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <assert.h>
#include <omp.h>
diff --git a/offload/test/api/omp_get_num_devices.c b/offload/test/api/omp_get_num_devices.c
index e8fbdcd1a1de0..97bd5b4927410 100644
--- a/offload/test/api/omp_get_num_devices.c
+++ b/offload/test/api/omp_get_num_devices.c
@@ -1,4 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/api/omp_get_num_procs.c b/offload/test/api/omp_get_num_procs.c
index 24222d32983b4..a0b9a0eaf4a4f 100644
--- a/offload/test/api/omp_get_num_procs.c
+++ b/offload/test/api/omp_get_num_procs.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/api/omp_indirect_call.c b/offload/test/api/omp_indirect_call.c
index b3e98d56a7706..c74cdfc903066 100644
--- a/offload/test/api/omp_indirect_call.c
+++ b/offload/test/api/omp_indirect_call.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <assert.h>
#include <stdio.h>
diff --git a/offload/test/api/ompx_3d.c b/offload/test/api/ompx_3d.c
index 2ae17c226f3a4..3dc902fe765d2 100644
--- a/offload/test/api/ompx_3d.c
+++ b/offload/test/api/ompx_3d.c
@@ -1,4 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <ompx.h>
diff --git a/offload/test/api/ompx_3d.cpp b/offload/test/api/ompx_3d.cpp
index 5b5491263a917..75dec5fbf2f8e 100644
--- a/offload/test/api/ompx_3d.cpp
+++ b/offload/test/api/ompx_3d.cpp
@@ -1,4 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <ompx.h>
diff --git a/offload/test/mapping/array_section_implicit_capture.c b/offload/test/mapping/array_section_implicit_capture.c
index fae18112763e5..c2b80b201181a 100644
--- a/offload/test/mapping/array_section_implicit_capture.c
+++ b/offload/test/mapping/array_section_implicit_capture.c
@@ -1,7 +1,8 @@
// RUN: %libomptarget-compile-generic
// RUN: %libomptarget-run-generic 2>&1 \
// RUN: | %fcheck-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/mapping/chained_containing_structs_1.cc b/offload/test/mapping/chained_containing_structs_1.cc
index e98d7fe1f9758..dd4e26c1195ff 100644
--- a/offload/test/mapping/chained_containing_structs_1.cc
+++ b/offload/test/mapping/chained_containing_structs_1.cc
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <cstdlib>
#include <cstdio>
diff --git a/offload/test/mapping/chained_containing_structs_2.cc b/offload/test/mapping/chained_containing_structs_2.cc
index ed959dd6c4958..247a75a8b569b 100644
--- a/offload/test/mapping/chained_containing_structs_2.cc
+++ b/offload/test/mapping/chained_containing_structs_2.cc
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <cstdlib>
#include <cstdio>
diff --git a/offload/test/mapping/chained_containing_structs_3.cc b/offload/test/mapping/chained_containing_structs_3.cc
index 95692e67a83bf..1c23b0e5a1755 100644
--- a/offload/test/mapping/chained_containing_structs_3.cc
+++ b/offload/test/mapping/chained_containing_structs_3.cc
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <cstdlib>
#include <cstdio>
diff --git a/offload/test/mapping/data_member_ref.cpp b/offload/test/mapping/data_member_ref.cpp
index 2b5f978b60404..b284706309adc 100644
--- a/offload/test/mapping/data_member_ref.cpp
+++ b/offload/test/mapping/data_member_ref.cpp
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/declare_mapper_nested_default_mappers.cpp b/offload/test/mapping/declare_mapper_nested_default_mappers.cpp
index 2d69f9573ad24..e9df27243a724 100644
--- a/offload/test/mapping/declare_mapper_nested_default_mappers.cpp
+++ b/offload/test/mapping/declare_mapper_nested_default_mappers.cpp
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <cstdio>
#include <cstdlib>
diff --git a/offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp b/offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp
index f822d08ba1e44..f337a593af828 100644
--- a/offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp
+++ b/offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp
@@ -1,7 +1,8 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
// UNSUPPORTED: amdgcn-amd-amdhsa
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
extern "C" int printf(const char *, ...);
diff --git a/offload/test/mapping/declare_mapper_nested_mappers.cpp b/offload/test/mapping/declare_mapper_nested_mappers.cpp
index 5d18739bf22ee..90c6d84ddab70 100644
--- a/offload/test/mapping/declare_mapper_nested_mappers.cpp
+++ b/offload/test/mapping/declare_mapper_nested_mappers.cpp
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <cstdio>
#include <cstdlib>
diff --git a/offload/test/mapping/declare_mapper_target_checks.cpp b/offload/test/mapping/declare_mapper_target_checks.cpp
index 562e283d81bcb..8f2f6c01419ab 100644
--- a/offload/test/mapping/declare_mapper_target_checks.cpp
+++ b/offload/test/mapping/declare_mapper_target_checks.cpp
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/delete_inf_refcount.c b/offload/test/mapping/delete_inf_refcount.c
index f6016c19b0ffb..dd7bd7160e8ae 100644
--- a/offload/test/mapping/delete_inf_refcount.c
+++ b/offload/test/mapping/delete_inf_refcount.c
@@ -1,4 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/duplicate_mappings_1.cpp b/offload/test/mapping/duplicate_mappings_1.cpp
index 731bff9035d5a..9b77838d9ccdd 100644
--- a/offload/test/mapping/duplicate_mappings_1.cpp
+++ b/offload/test/mapping/duplicate_mappings_1.cpp
@@ -1,6 +1,7 @@
// clang-format off
// RUN: %libomptarget-compilexx-generic -Wno-openmp-mapping && %libomptarget-run-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// clang-format on
diff --git a/offload/test/mapping/duplicate_mappings_2.cpp b/offload/test/mapping/duplicate_mappings_2.cpp
index 6ca7d3b55dfb5..8122c48bd6be9 100644
--- a/offload/test/mapping/duplicate_mappings_2.cpp
+++ b/offload/test/mapping/duplicate_mappings_2.cpp
@@ -1,6 +1,7 @@
// clang-format off
// RUN: %libomptarget-compilexx-generic -Wno-openmp-mapping && %libomptarget-run-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <assert.h>
diff --git a/offload/test/mapping/firstprivate_aligned.cpp b/offload/test/mapping/firstprivate_aligned.cpp
index 4521f4397112b..955e4ed708646 100644
--- a/offload/test/mapping/firstprivate_aligned.cpp
+++ b/offload/test/mapping/firstprivate_aligned.cpp
@@ -1,4 +1,5 @@
// RUN: %libomptarget-compilexx-generic -O3 && %libomptarget-run-generic
+// https://github.com/llvm/llvm-project/issues/182119
// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/implicit_device_ptr.c b/offload/test/mapping/implicit_device_ptr.c
index 6e32ece82a1c4..334265a1f8a82 100644
--- a/offload/test/mapping/implicit_device_ptr.c
+++ b/offload/test/mapping/implicit_device_ptr.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/low_alignment.c b/offload/test/mapping/low_alignment.c
index 71e90b3695a35..3f2cceea1700a 100644
--- a/offload/test/mapping/low_alignment.c
+++ b/offload/test/mapping/low_alignment.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/map_both_pointer_pointee.c b/offload/test/mapping/map_both_pointer_pointee.c
index 0462beaf184df..85326f6680d7b 100644
--- a/offload/test/mapping/map_both_pointer_pointee.c
+++ b/offload/test/mapping/map_both_pointer_pointee.c
@@ -3,7 +3,8 @@
// REQUIRES: unified_shared_memory
// UNSUPPORTED: amdgcn-amd-amdhsa
//
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#pragma omp declare target
int *ptr1;
diff --git a/offload/test/mapping/map_ordering_tgt_alloc_tofrom.c b/offload/test/mapping/map_ordering_tgt_alloc_tofrom.c
index af7463066a80b..3055722178ecc 100644
--- a/offload/test/mapping/map_ordering_tgt_alloc_tofrom.c
+++ b/offload/test/mapping/map_ordering_tgt_alloc_tofrom.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/map_ordering_tgt_data_alloc_from.c b/offload/test/mapping/map_ordering_tgt_data_alloc_from.c
index eaeb38df9c651..2a98bcd0bd59f 100644
--- a/offload/test/mapping/map_ordering_tgt_data_alloc_from.c
+++ b/offload/test/mapping/map_ordering_tgt_data_alloc_from.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/map_ordering_tgt_data_alloc_to_from.c b/offload/test/mapping/map_ordering_tgt_data_alloc_to_from.c
index 2e501ecaa8bd9..aa5a3b02535dd 100644
--- a/offload/test/mapping/map_ordering_tgt_data_alloc_to_from.c
+++ b/offload/test/mapping/map_ordering_tgt_data_alloc_to_from.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/map_ordering_tgt_data_alloc_tofrom.c b/offload/test/mapping/map_ordering_tgt_data_alloc_tofrom.c
index 883a59d0d4082..f22303480ece1 100644
--- a/offload/test/mapping/map_ordering_tgt_data_alloc_tofrom.c
+++ b/offload/test/mapping/map_ordering_tgt_data_alloc_tofrom.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/map_ordering_tgt_exit_data_delete_from.c b/offload/test/mapping/map_ordering_tgt_exit_data_delete_from.c
index d685d99c350a0..6455686a0f4a1 100644
--- a/offload/test/mapping/map_ordering_tgt_exit_data_delete_from.c
+++ b/offload/test/mapping/map_ordering_tgt_exit_data_delete_from.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/ompx_hold/struct.c b/offload/test/mapping/ompx_hold/struct.c
index 0b898d5d3b567..d46f4bbc68ccf 100644
--- a/offload/test/mapping/ompx_hold/struct.c
+++ b/offload/test/mapping/ompx_hold/struct.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-generic -fopenmp-extensions
// RUN: %libomptarget-run-generic | %fcheck-generic -strict-whitespace
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/ompx_hold/target.c b/offload/test/mapping/ompx_hold/target.c
index 40fd14c6bf005..e17e72cf07028 100644
--- a/offload/test/mapping/ompx_hold/target.c
+++ b/offload/test/mapping/ompx_hold/target.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-generic -fopenmp-extensions
// RUN: %libomptarget-run-generic | %fcheck-generic -strict-whitespace
-
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/pr38704.c b/offload/test/mapping/pr38704.c
index 612c2683b37ce..411eeaa12a6c7 100644
--- a/offload/test/mapping/pr38704.c
+++ b/offload/test/mapping/pr38704.c
@@ -3,7 +3,8 @@
// Clang 6.0 doesn't use the new map interface, undefined behavior when
// the compiler emits "old" interface code for structures.
// UNSUPPORTED: clang-6
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/mapping/present/target.c b/offload/test/mapping/present/target.c
index 4344c42c808ff..63a9f9e0cdc3b 100644
--- a/offload/test/mapping/present/target.c
+++ b/offload/test/mapping/present/target.c
@@ -1,6 +1,8 @@
// RUN: %libomptarget-compile-generic
// RUN: %libomptarget-run-fail-generic 2>&1 \
// RUN: | %fcheck-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/present/target_array_extension.c b/offload/test/mapping/present/target_array_extension.c
index 873b2b36170ee..876c39aa0de01 100644
--- a/offload/test/mapping/present/target_array_extension.c
+++ b/offload/test/mapping/present/target_array_extension.c
@@ -16,6 +16,8 @@
// RUN: %libomptarget-run-fail-generic 2>&1 \
// RUN: | %fcheck-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// END.
#include <stdio.h>
diff --git a/offload/test/mapping/present/unified_shared_memory.c b/offload/test/mapping/present/unified_shared_memory.c
index ab6e3bd0e5fa7..06007615f1776 100644
--- a/offload/test/mapping/present/unified_shared_memory.c
+++ b/offload/test/mapping/present/unified_shared_memory.c
@@ -3,6 +3,8 @@
// RUN: | %fcheck-generic
// REQUIRES: unified_shared_memory
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/present/zero_length_array_section.c b/offload/test/mapping/present/zero_length_array_section.c
index e903a268ef335..1a99617ed0165 100644
--- a/offload/test/mapping/present/zero_length_array_section.c
+++ b/offload/test/mapping/present/zero_length_array_section.c
@@ -1,6 +1,8 @@
// RUN: %libomptarget-compile-generic
// RUN: %libomptarget-run-fail-generic 2>&1 \
// RUN: | %fcheck-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/present/zero_length_array_section_exit.c b/offload/test/mapping/present/zero_length_array_section_exit.c
index 5a7360542ec3c..6d11e3accb6fb 100644
--- a/offload/test/mapping/present/zero_length_array_section_exit.c
+++ b/offload/test/mapping/present/zero_length_array_section_exit.c
@@ -1,6 +1,8 @@
// RUN: %libomptarget-compile-generic
// RUN: %libomptarget-run-fail-generic 2>&1 \
// RUN: | %fcheck-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/ptr_and_obj_motion.c b/offload/test/mapping/ptr_and_obj_motion.c
index 6a5f16eb8f301..d58c28a5fd4c2 100644
--- a/offload/test/mapping/ptr_and_obj_motion.c
+++ b/offload/test/mapping/ptr_and_obj_motion.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/target_data_array_extension_at_exit.c b/offload/test/mapping/target_data_array_extension_at_exit.c
index cfb09a98e06f1..9cff8f8dabe0e 100644
--- a/offload/test/mapping/target_data_array_extension_at_exit.c
+++ b/offload/test/mapping/target_data_array_extension_at_exit.c
@@ -15,7 +15,8 @@
// RUN: -DEXTENDS=AFTER
// RUN: %libomptarget-run-generic 2>&1 \
// RUN: | %fcheck-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// END.
diff --git a/offload/test/mapping/target_derefence_array_pointrs.cpp b/offload/test/mapping/target_derefence_array_pointrs.cpp
index ed3c1ed17b243..4c464454a58cb 100644
--- a/offload/test/mapping/target_derefence_array_pointrs.cpp
+++ b/offload/test/mapping/target_derefence_array_pointrs.cpp
@@ -6,7 +6,8 @@
// UNSUPPORTED: amdgcn-amd-amdhsa
// UNSUPPORTED: nvptx64-nvidia-cuda
// UNSUPPORTED: nvptx64-nvidia-cuda-LTO
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/mapping/target_has_device_addr.c b/offload/test/mapping/target_has_device_addr.c
index 5bff5168d4471..21f373ad8dec6 100644
--- a/offload/test/mapping/target_has_device_addr.c
+++ b/offload/test/mapping/target_has_device_addr.c
@@ -3,7 +3,8 @@
// RUN: | %fcheck-generic
// UNSUPPORTED: amdgcn-amd-amdhsa
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/target_implicit_partial_map.c b/offload/test/mapping/target_implicit_partial_map.c
index a8b2cc893df20..a9d40537916e1 100644
--- a/offload/test/mapping/target_implicit_partial_map.c
+++ b/offload/test/mapping/target_implicit_partial_map.c
@@ -1,7 +1,8 @@
// RUN: %libomptarget-compile-generic
// RUN: %libomptarget-run-generic 2>&1 \
// RUN: | %fcheck-generic
-
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// END.
#include <omp.h>
diff --git a/offload/test/mapping/target_map_for_member_data.cpp b/offload/test/mapping/target_map_for_member_data.cpp
index 9f2adc29d6671..5a31d546efd8f 100644
--- a/offload/test/mapping/target_map_for_member_data.cpp
+++ b/offload/test/mapping/target_map_for_member_data.cpp
@@ -1,7 +1,8 @@
// RUN: %libomptarget-compile-generic -fopenmp-version=51
// RUN: %libomptarget-run-generic 2>&1 \
// RUN: | %fcheck-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
extern "C" int printf(const char *, ...);
template <typename T> class A {
diff --git a/offload/test/offloading/atomic-compare-signedness.c b/offload/test/offloading/atomic-compare-signedness.c
index 2fd6a484460d0..7ce981f463837 100644
--- a/offload/test/offloading/atomic-compare-signedness.c
+++ b/offload/test/offloading/atomic-compare-signedness.c
@@ -7,6 +7,7 @@
// RUN: %libomptarget-run-generic | %fcheck-generic
// RUN: %libomptarget-compileopt-generic -fopenmp-version=51
// RUN: %libomptarget-run-generic | %fcheck-generic
+// https://github.com/llvm/llvm-project/issues/182119
// UNSUPPORTED: intelgpu
// High parallelism increases our chances of detecting a lack of atomicity.
diff --git a/offload/test/offloading/bug47654.cpp b/offload/test/offloading/bug47654.cpp
index 3203e9f6da489..144321307536f 100644
--- a/offload/test/offloading/bug47654.cpp
+++ b/offload/test/offloading/bug47654.cpp
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
// RUN: %libomptarget-compileoptxx-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <cassert>
#include <iostream>
diff --git a/offload/test/offloading/bug49021.cpp b/offload/test/offloading/bug49021.cpp
index 6d919871f0d26..83fc66317e9be 100644
--- a/offload/test/offloading/bug49021.cpp
+++ b/offload/test/offloading/bug49021.cpp
@@ -3,6 +3,7 @@
// RUN: %libomptarget-compilexx-generic -O3 -ffast-math && %libomptarget-run-generic
// RUN: %libomptarget-compileoptxx-generic -O3 && %libomptarget-run-generic
// RUN: %libomptarget-compileoptxx-generic -O3 -ffast-math && %libomptarget-run-generic
+// https://github.com/llvm/llvm-project/issues/182119
// UNSUPPORTED: intelgpu
// clang-format on
diff --git a/offload/test/offloading/bug50022.cpp b/offload/test/offloading/bug50022.cpp
index 1bd48be9f9d88..ad6d9ece01782 100644
--- a/offload/test/offloading/bug50022.cpp
+++ b/offload/test/offloading/bug50022.cpp
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compilexx-and-run-generic
// RUN: %libomptarget-compileoptxx-and-run-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <cassert>
#include <iostream>
diff --git a/offload/test/offloading/bug53727.cpp b/offload/test/offloading/bug53727.cpp
index a1a01f0078508..9a81b530362e5 100644
--- a/offload/test/offloading/bug53727.cpp
+++ b/offload/test/offloading/bug53727.cpp
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compilexx-and-run-generic
// RUN: %libomptarget-compileoptxx-and-run-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <cassert>
#include <iostream>
diff --git a/offload/test/offloading/bug64959.c b/offload/test/offloading/bug64959.c
index eddc55325ffe9..6b563f09087fd 100644
--- a/offload/test/offloading/bug64959.c
+++ b/offload/test/offloading/bug64959.c
@@ -6,6 +6,8 @@
// UNSUPPORTED: amdgcn-amd-amdhsa
// UNSUPPORTED: nvptx64-nvidia-cuda
// UNSUPPORTED: nvptx64-nvidia-cuda-LTO
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/ctor_dtor.cpp b/offload/test/offloading/ctor_dtor.cpp
index d8ec93c7ae479..74561e4e1c516 100644
--- a/offload/test/offloading/ctor_dtor.cpp
+++ b/offload/test/offloading/ctor_dtor.cpp
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
// RUN: %libomptarget-compileoptxx-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <cstdio>
struct S {
diff --git a/offload/test/offloading/dynamic_module_load.c b/offload/test/offloading/dynamic_module_load.c
index 2e49cd35e5328..9316c96bf5770 100644
--- a/offload/test/offloading/dynamic_module_load.c
+++ b/offload/test/offloading/dynamic_module_load.c
@@ -1,6 +1,8 @@
// clang-format off
// RUN: %libomptarget-compile-generic -DSHARED -fPIC -shared -o %t.so && %clang %flags %s -o %t -ldl && %libomptarget-run-generic %t.so 2>&1 | %fcheck-generic
// clang-format on
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#ifdef SHARED
#include <stdio.h>
int foo() {
diff --git a/offload/test/offloading/extern.c b/offload/test/offloading/extern.c
index f594e377d8f24..0a1af3367de78 100644
--- a/offload/test/offloading/extern.c
+++ b/offload/test/offloading/extern.c
@@ -1,7 +1,8 @@
// clang-format off
// RUN: %libomptarget-compile-generic -DVAR -c -o %t.o
// RUN: %libomptarget-compile-generic %t.o && %libomptarget-run-generic | %fcheck-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// clang-format on
#ifdef VAR
int x = 1;
diff --git a/offload/test/offloading/force-usm.cpp b/offload/test/offloading/force-usm.cpp
index 9afe57d797356..acb8a6d5ad070 100644
--- a/offload/test/offloading/force-usm.cpp
+++ b/offload/test/offloading/force-usm.cpp
@@ -10,7 +10,8 @@
//
// UNSUPPORTED: nvptx64-nvidia-cuda
// UNSUPPORTED: nvptx64-nvidia-cuda-LTO
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// clang-format on
#include <cassert>
diff --git a/offload/test/offloading/global_constructor.cpp b/offload/test/offloading/global_constructor.cpp
index 0f869788e146f..4892e50f572ca 100644
--- a/offload/test/offloading/global_constructor.cpp
+++ b/offload/test/offloading/global_constructor.cpp
@@ -1,6 +1,7 @@
// clang-format off
// RUN: %libomptarget-compilexx-generic && %libomptarget-run-generic | %fcheck-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// clang-format on
#include <cstdio>
diff --git a/offload/test/offloading/host_as_target.c b/offload/test/offloading/host_as_target.c
index 24e870b441edb..a806bd181c01f 100644
--- a/offload/test/offloading/host_as_target.c
+++ b/offload/test/offloading/host_as_target.c
@@ -6,7 +6,8 @@
// - Works whether it's specified directly or as the default device.
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/info.c b/offload/test/offloading/info.c
index d86644b871e25..cebe99d093863 100644
--- a/offload/test/offloading/info.c
+++ b/offload/test/offloading/info.c
@@ -7,6 +7,8 @@
// FIXME: Fails due to optimized debugging in 'ptxas'.
// UNSUPPORTED: nvptx64-nvidia-cuda-LTO
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/memory_manager.cpp b/offload/test/offloading/memory_manager.cpp
index d6d8697fcdec8..811fe590507fd 100644
--- a/offload/test/offloading/memory_manager.cpp
+++ b/offload/test/offloading/memory_manager.cpp
@@ -1,4 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
diff --git a/offload/test/offloading/offloading_success.c b/offload/test/offloading/offloading_success.c
index f849cb4cfba79..89f33f6ac233f 100644
--- a/offload/test/offloading/offloading_success.c
+++ b/offload/test/offloading/offloading_success.c
@@ -1,4 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/offloading_success.cpp b/offload/test/offloading/offloading_success.cpp
index d3103b8666ab5..41aefacaa8136 100644
--- a/offload/test/offloading/offloading_success.cpp
+++ b/offload/test/offloading/offloading_success.cpp
@@ -1,4 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/parallel_offloading_map.cpp b/offload/test/offloading/parallel_offloading_map.cpp
index d2a6c4200df52..fe61936508b1a 100644
--- a/offload/test/offloading/parallel_offloading_map.cpp
+++ b/offload/test/offloading/parallel_offloading_map.cpp
@@ -1,6 +1,8 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
// REQUIRES: gpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <cassert>
#include <iostream>
diff --git a/offload/test/offloading/requires.c b/offload/test/offloading/requires.c
index fc21ca3420e9e..188ed2d03f5ff 100644
--- a/offload/test/offloading/requires.c
+++ b/offload/test/offloading/requires.c
@@ -1,6 +1,8 @@
// clang-format off
// RUN: %libomptarget-compile-generic -DREQ=1 && %libomptarget-run-generic 2>&1 | %fcheck-generic -check-prefix=GOOD
// RUN: %libomptarget-compile-generic -DREQ=2 && %not --crash %libomptarget-run-generic 2>&1 | %fcheck-generic -check-prefix=BAD
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// clang-format on
/*
diff --git a/offload/test/offloading/static_linking.c b/offload/test/offloading/static_linking.c
index 000ed11052d37..d84ecbac74cd2 100644
--- a/offload/test/offloading/static_linking.c
+++ b/offload/test/offloading/static_linking.c
@@ -2,7 +2,8 @@
// RUN: %libomptarget-compile-generic -DLIBRARY -c -o %t.o
// RUN: ar rcs %t.a %t.o
// RUN: %libomptarget-compile-generic %t.a && %libomptarget-run-generic 2>&1 | %fcheck-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// clang-format on
#ifdef LIBRARY
diff --git a/offload/test/offloading/strided_multiple_update_from.c b/offload/test/offloading/strided_multiple_update_from.c
index 4f2df81933738..c048d0bcb21c4 100644
--- a/offload/test/offloading/strided_multiple_update_from.c
+++ b/offload/test/offloading/strided_multiple_update_from.c
@@ -3,7 +3,8 @@
// from the device to the host.
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_multiple_update_to.c b/offload/test/offloading/strided_multiple_update_to.c
index bb16d7a09269a..1964cbf3f37cc 100644
--- a/offload/test/offloading/strided_multiple_update_to.c
+++ b/offload/test/offloading/strided_multiple_update_to.c
@@ -3,7 +3,8 @@
// from the host to the device.
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_partial_update_from.c b/offload/test/offloading/strided_partial_update_from.c
index 4a2977a08f70b..14ef6a882fee5 100644
--- a/offload/test/offloading/strided_partial_update_from.c
+++ b/offload/test/offloading/strided_partial_update_from.c
@@ -3,7 +3,8 @@
// across the array
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_partial_update_to.c b/offload/test/offloading/strided_partial_update_to.c
index f9c960fb9c965..bc53ab6f52d46 100644
--- a/offload/test/offloading/strided_partial_update_to.c
+++ b/offload/test/offloading/strided_partial_update_to.c
@@ -3,7 +3,8 @@
// across the array
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_ptr_multiple_update_from.c b/offload/test/offloading/strided_ptr_multiple_update_from.c
index 33deb59534f69..e61b810cbc13d 100644
--- a/offload/test/offloading/strided_ptr_multiple_update_from.c
+++ b/offload/test/offloading/strided_ptr_multiple_update_from.c
@@ -3,7 +3,8 @@
// from the device to the host using dynamically allocated memory.
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/offloading/strided_ptr_multiple_update_to.c b/offload/test/offloading/strided_ptr_multiple_update_to.c
index be56f3dc5c248..7fcf9c3f1267e 100644
--- a/offload/test/offloading/strided_ptr_multiple_update_to.c
+++ b/offload/test/offloading/strided_ptr_multiple_update_to.c
@@ -3,7 +3,8 @@
// from the host to the device using dynamically allocated memory.
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/offloading/strided_ptr_partial_update_from.c b/offload/test/offloading/strided_ptr_partial_update_from.c
index be235d3d9c532..c20f312a71b3c 100644
--- a/offload/test/offloading/strided_ptr_partial_update_from.c
+++ b/offload/test/offloading/strided_ptr_partial_update_from.c
@@ -3,7 +3,8 @@
// across the array using dynamically allocated memory.
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/offloading/strided_ptr_partial_update_to.c b/offload/test/offloading/strided_ptr_partial_update_to.c
index 662e1a2655cab..850f0f2e734a6 100644
--- a/offload/test/offloading/strided_ptr_partial_update_to.c
+++ b/offload/test/offloading/strided_ptr_partial_update_to.c
@@ -3,7 +3,8 @@
// across the array using dynamically allocated memory.
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/offloading/strided_update_count_expression.c b/offload/test/offloading/strided_update_count_expression.c
index e18fc881a8021..f005fde7dd4da 100644
--- a/offload/test/offloading/strided_update_count_expression.c
+++ b/offload/test/offloading/strided_update_count_expression.c
@@ -4,6 +4,8 @@
// where the count (len/2) is a variable expression, not a constant.
// RUN: %libomptarget-compile-run-and-check-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_count_expression_complex.c b/offload/test/offloading/strided_update_count_expression_complex.c
index 560c4e767f882..288f93df0ff8a 100644
--- a/offload/test/offloading/strided_update_count_expression_complex.c
+++ b/offload/test/offloading/strided_update_count_expression_complex.c
@@ -1,7 +1,8 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Tests non-contiguous array sections with complex expression-based count
// scenarios including multiple struct arrays and non-zero offset.
-
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_from.c b/offload/test/offloading/strided_update_from.c
index 9910bed9c5466..779834d6d37eb 100644
--- a/offload/test/offloading/strided_update_from.c
+++ b/offload/test/offloading/strided_update_from.c
@@ -4,7 +4,8 @@
// other element (stride 2) from the device to the host
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_multiple_arrays_count_expression.c b/offload/test/offloading/strided_update_multiple_arrays_count_expression.c
index a1472cacc4a39..8490b262dc47b 100644
--- a/offload/test/offloading/strided_update_multiple_arrays_count_expression.c
+++ b/offload/test/offloading/strided_update_multiple_arrays_count_expression.c
@@ -4,6 +4,8 @@
// same array with various count expressions.
// RUN: %libomptarget-compile-run-and-check-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_multiple_arrays_variable_stride.c b/offload/test/offloading/strided_update_multiple_arrays_variable_stride.c
index 32859202a20fa..01c58d53ac9d6 100644
--- a/offload/test/offloading/strided_update_multiple_arrays_variable_stride.c
+++ b/offload/test/offloading/strided_update_multiple_arrays_variable_stride.c
@@ -1,7 +1,8 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Tests multiple arrays with different variable strides in single update
// clause.
-
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_to.c b/offload/test/offloading/strided_update_to.c
index eca20abed4b8d..0683f358f31f6 100644
--- a/offload/test/offloading/strided_update_to.c
+++ b/offload/test/offloading/strided_update_to.c
@@ -4,7 +4,8 @@
// other element (stride 2) from the host to the device
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_variable_count_and_stride.c b/offload/test/offloading/strided_update_variable_count_and_stride.c
index 1e1e41653c2c4..28597b9e4ee0e 100644
--- a/offload/test/offloading/strided_update_variable_count_and_stride.c
+++ b/offload/test/offloading/strided_update_variable_count_and_stride.c
@@ -1,7 +1,8 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Tests combining variable count expression AND variable stride in array
// sections.
-
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_variable_stride.c b/offload/test/offloading/strided_update_variable_stride.c
index 7c8079efa56ea..0cee4c832cefa 100644
--- a/offload/test/offloading/strided_update_variable_stride.c
+++ b/offload/test/offloading/strided_update_variable_stride.c
@@ -2,6 +2,8 @@
// Tests data[0:5:stride] where stride is a variable, making it non-contiguous.
// RUN: %libomptarget-compile-run-and-check-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_variable_stride_misc.c b/offload/test/offloading/strided_update_variable_stride_misc.c
index df2bfa64cfe3c..1220027be27f5 100644
--- a/offload/test/offloading/strided_update_variable_stride_misc.c
+++ b/offload/test/offloading/strided_update_variable_stride_misc.c
@@ -1,6 +1,8 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Miscellaneous variable stride tests: stride=1, stride=array_size, stride from
// array subscript.
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target-teams-atomic.c b/offload/test/offloading/target-teams-atomic.c
index 97563422ced8f..f9b554438c2ba 100644
--- a/offload/test/offloading/target-teams-atomic.c
+++ b/offload/test/offloading/target-teams-atomic.c
@@ -3,6 +3,8 @@
// default.
// RUN: %libomptarget-compile-run-and-check-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdbool.h>
diff --git a/offload/test/offloading/target-tile.c b/offload/test/offloading/target-tile.c
index e5413a9493f4c..e29da170cf940 100644
--- a/offload/test/offloading/target-tile.c
+++ b/offload/test/offloading/target-tile.c
@@ -3,7 +3,8 @@
// RUN: %libomptarget-compile-generic -fopenmp-version=51
// RUN: %libomptarget-run-generic 2>&1 | %fcheck-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/offloading/target_depend_nowait.cpp b/offload/test/offloading/target_depend_nowait.cpp
index 0a6b903db6882..fe060aac6ad95 100644
--- a/offload/test/offloading/target_depend_nowait.cpp
+++ b/offload/test/offloading/target_depend_nowait.cpp
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_nowait_target.cpp b/offload/test/offloading/target_nowait_target.cpp
index f6c2b56963cf1..a8cafaec46dbf 100644
--- a/offload/test/offloading/target_nowait_target.cpp
+++ b/offload/test/offloading/target_nowait_target.cpp
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compilexx-and-run-generic
// RUN: %libomptarget-compileoptxx-and-run-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <cassert>
diff --git a/offload/test/offloading/target_update_from.c b/offload/test/offloading/target_update_from.c
index 92697bc6b802a..b10568a08f04d 100644
--- a/offload/test/offloading/target_update_from.c
+++ b/offload/test/offloading/target_update_from.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// This test checks that "update from" clause in OpenMP supports strided
// sections. #pragma omp target update from(result[0:N/2:2]) updates every other
// element from device
diff --git a/offload/test/offloading/target_update_ptr_count_expression.c b/offload/test/offloading/target_update_ptr_count_expression.c
index c4b9fd566d401..e23684be51bc3 100644
--- a/offload/test/offloading/target_update_ptr_count_expression.c
+++ b/offload/test/offloading/target_update_ptr_count_expression.c
@@ -1,6 +1,8 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Tests non-contiguous array sections with expression-based count on
// heap-allocated pointer arrays with both FROM and TO directives.
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_update_ptr_variable_count_and_stride.c b/offload/test/offloading/target_update_ptr_variable_count_and_stride.c
index 1a28595969c69..74214a2225189 100644
--- a/offload/test/offloading/target_update_ptr_variable_count_and_stride.c
+++ b/offload/test/offloading/target_update_ptr_variable_count_and_stride.c
@@ -1,6 +1,8 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Tests heap-allocated pointers with both variable count expression and
// variable stride.
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_update_ptr_variable_stride.c b/offload/test/offloading/target_update_ptr_variable_stride.c
index bea396065b760..dbe7d1414d981 100644
--- a/offload/test/offloading/target_update_ptr_variable_stride.c
+++ b/offload/test/offloading/target_update_ptr_variable_stride.c
@@ -1,6 +1,8 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Tests non-contiguous array sections with variable stride on heap-allocated
// pointers.
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_update_strided_struct_count_expression.c b/offload/test/offloading/target_update_strided_struct_count_expression.c
index 89b0e1b4c7ea7..1af9ff045a3f2 100644
--- a/offload/test/offloading/target_update_strided_struct_count_expression.c
+++ b/offload/test/offloading/target_update_strided_struct_count_expression.c
@@ -1,6 +1,8 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Tests non-contiguous array sections with expression-based count on struct
// member arrays with both FROM and TO directives.
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_update_strided_struct_from.c b/offload/test/offloading/target_update_strided_struct_from.c
index 7a8805ab35e98..9f13e61504a9a 100644
--- a/offload/test/offloading/target_update_strided_struct_from.c
+++ b/offload/test/offloading/target_update_strided_struct_from.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// This test checks that "update from" with user-defined mapper supports strided
// sections using fixed-size arrays in structs.
diff --git a/offload/test/offloading/target_update_strided_struct_multiple_from.c b/offload/test/offloading/target_update_strided_struct_multiple_from.c
index 16023be5970a1..ae82ba04f9be0 100644
--- a/offload/test/offloading/target_update_strided_struct_multiple_from.c
+++ b/offload/test/offloading/target_update_strided_struct_multiple_from.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// This test checks that #pragma omp target update from(s1.data[0:6:2],
// s2.data[0:4:3]) correctly updates strided sections covering the full arrays
// from device to host.
diff --git a/offload/test/offloading/target_update_strided_struct_multiple_to.c b/offload/test/offloading/target_update_strided_struct_multiple_to.c
index 182b40cb06b37..a61268f4e19e2 100644
--- a/offload/test/offloading/target_update_strided_struct_multiple_to.c
+++ b/offload/test/offloading/target_update_strided_struct_multiple_to.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// This test checks that #pragma omp target update to(s1.data[0:6:2],
// s2.data[0:4:3]) correctly updates strided sections covering the full arrays
// from host to device.
diff --git a/offload/test/offloading/target_update_strided_struct_partial_from.c b/offload/test/offloading/target_update_strided_struct_partial_from.c
index fb852c07f0238..4f98a20f90373 100644
--- a/offload/test/offloading/target_update_strided_struct_partial_from.c
+++ b/offload/test/offloading/target_update_strided_struct_partial_from.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// This test checks that #pragma omp target update from(s.data[0:2:3]) correctly
// updates every third element (stride 3) from the device to the host
// using a struct with fixed-size array member.
diff --git a/offload/test/offloading/target_update_strided_struct_partial_to.c b/offload/test/offloading/target_update_strided_struct_partial_to.c
index 35c52ea2db207..baff8a45b2dfd 100644
--- a/offload/test/offloading/target_update_strided_struct_partial_to.c
+++ b/offload/test/offloading/target_update_strided_struct_partial_to.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// This test checks that #pragma omp target update to(s.data[0:2:3]) correctly
// updates every third element (stride 3) from the host to the device
// for struct member arrays.
diff --git a/offload/test/offloading/target_update_strided_struct_to.c b/offload/test/offloading/target_update_strided_struct_to.c
index 3c02232df2806..b323740504fd1 100644
--- a/offload/test/offloading/target_update_strided_struct_to.c
+++ b/offload/test/offloading/target_update_strided_struct_to.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// This test checks that "update to" with struct member arrays supports strided
// sections using fixed-size arrays in structs.
diff --git a/offload/test/offloading/target_update_strided_struct_variable_count_and_stride.c b/offload/test/offloading/target_update_strided_struct_variable_count_and_stride.c
index 6daf10383e921..0fb7f70ad7546 100644
--- a/offload/test/offloading/target_update_strided_struct_variable_count_and_stride.c
+++ b/offload/test/offloading/target_update_strided_struct_variable_count_and_stride.c
@@ -1,6 +1,8 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Tests struct member arrays with both variable count expression and variable
// stride.
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_update_strided_struct_variable_stride.c b/offload/test/offloading/target_update_strided_struct_variable_stride.c
index 4cd9da629ca93..bb68b7f323bc6 100644
--- a/offload/test/offloading/target_update_strided_struct_variable_stride.c
+++ b/offload/test/offloading/target_update_strided_struct_variable_stride.c
@@ -1,6 +1,8 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Tests non-contiguous array sections with variable stride on struct member
// arrays.
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_update_to.c b/offload/test/offloading/target_update_to.c
index ca4984850a493..043b18018b951 100644
--- a/offload/test/offloading/target_update_to.c
+++ b/offload/test/offloading/target_update_to.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// This test checks that "update to" clause in OpenMP supports strided sections.
// #pragma omp target update to(result[0:8:2]) updates every other element
// (stride 2)
diff --git a/offload/test/offloading/task_in_reduction_target.c b/offload/test/offloading/task_in_reduction_target.c
index b546d73d544a5..9281294128792 100644
--- a/offload/test/offloading/task_in_reduction_target.c
+++ b/offload/test/offloading/task_in_reduction_target.c
@@ -1,5 +1,7 @@
// RUN: %libomptarget-compile-generic && \
// RUN: %libomptarget-run-generic
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/workshare_chunk.c b/offload/test/offloading/workshare_chunk.c
index 551cc8d997cf7..4e8525282504a 100644
--- a/offload/test/offloading/workshare_chunk.c
+++ b/offload/test/offloading/workshare_chunk.c
@@ -2,7 +2,8 @@
// RUN: %libomptarget-compileopt-run-and-check-generic
// REQUIRES: gpu
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// clang-format off
diff --git a/offload/test/sanitizer/kernel_crash.c b/offload/test/sanitizer/kernel_crash.c
index 87133bfcfbd6e..a0d1ecab9aff0 100644
--- a/offload/test/sanitizer/kernel_crash.c
+++ b/offload/test/sanitizer/kernel_crash.c
@@ -12,7 +12,8 @@
// UNSUPPORTED: aarch64-unknown-linux-gnu
// UNSUPPORTED: x86_64-unknown-linux-gnu
// UNSUPPORTED: s390x-ibm-linux-gnu
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
diff --git a/offload/test/sanitizer/kernel_crash_async.c b/offload/test/sanitizer/kernel_crash_async.c
index 48bd65162db8d..73c34671acadd 100644
--- a/offload/test/sanitizer/kernel_crash_async.c
+++ b/offload/test/sanitizer/kernel_crash_async.c
@@ -12,7 +12,8 @@
// UNSUPPORTED: aarch64-unknown-linux-gnu
// UNSUPPORTED: x86_64-unknown-linux-gnu
// UNSUPPORTED: s390x-ibm-linux-gnu
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
diff --git a/offload/test/sanitizer/kernel_crash_many.c b/offload/test/sanitizer/kernel_crash_many.c
index 146b44586fa91..694b01792e8bb 100644
--- a/offload/test/sanitizer/kernel_crash_many.c
+++ b/offload/test/sanitizer/kernel_crash_many.c
@@ -10,7 +10,8 @@
// UNSUPPORTED: aarch64-unknown-linux-gnu
// UNSUPPORTED: x86_64-unknown-linux-gnu
// UNSUPPORTED: s390x-ibm-linux-gnu
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
diff --git a/offload/test/sanitizer/kernel_trap.c b/offload/test/sanitizer/kernel_trap.c
index 98d5faf241c85..3281bd2c8ca22 100644
--- a/offload/test/sanitizer/kernel_trap.c
+++ b/offload/test/sanitizer/kernel_trap.c
@@ -12,7 +12,8 @@
// UNSUPPORTED: aarch64-unknown-linux-gnu
// UNSUPPORTED: x86_64-unknown-linux-gnu
// UNSUPPORTED: s390x-ibm-linux-gnu
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
diff --git a/offload/test/sanitizer/kernel_trap.cpp b/offload/test/sanitizer/kernel_trap.cpp
index a83e6cbaa78a0..cb3a42a9daa47 100644
--- a/offload/test/sanitizer/kernel_trap.cpp
+++ b/offload/test/sanitizer/kernel_trap.cpp
@@ -13,7 +13,8 @@
// UNSUPPORTED: aarch64-unknown-linux-gnu
// UNSUPPORTED: x86_64-unknown-linux-gnu
// UNSUPPORTED: s390x-ibm-linux-gnu
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
struct S {};
diff --git a/offload/test/sanitizer/kernel_trap_async.c b/offload/test/sanitizer/kernel_trap_async.c
index aa31c6371fa0c..466fee11f5d00 100644
--- a/offload/test/sanitizer/kernel_trap_async.c
+++ b/offload/test/sanitizer/kernel_trap_async.c
@@ -12,6 +12,7 @@
// UNSUPPORTED: aarch64-unknown-linux-gnu
// UNSUPPORTED: x86_64-unknown-linux-gnu
// UNSUPPORTED: s390x-ibm-linux-gnu
+// https://github.com/llvm/llvm-project/issues/182119
// UNSUPPORTED: intelgpu
#include <omp.h>
diff --git a/offload/test/sanitizer/kernel_trap_many.c b/offload/test/sanitizer/kernel_trap_many.c
index c21cc99159694..72640728483fe 100644
--- a/offload/test/sanitizer/kernel_trap_many.c
+++ b/offload/test/sanitizer/kernel_trap_many.c
@@ -10,7 +10,8 @@
// UNSUPPORTED: aarch64-unknown-linux-gnu
// UNSUPPORTED: x86_64-unknown-linux-gnu
// UNSUPPORTED: s390x-ibm-linux-gnu
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
diff --git a/offload/test/sanitizer/use_after_free_1.c b/offload/test/sanitizer/use_after_free_1.c
index 2dee3d107a115..44a06f8506aa9 100644
--- a/offload/test/sanitizer/use_after_free_1.c
+++ b/offload/test/sanitizer/use_after_free_1.c
@@ -9,6 +9,7 @@
// UNSUPPORTED: nvidiagpu
//
// REQUIRES: gpu
+// https://github.com/llvm/llvm-project/issues/182119
// UNSUPPORTED: intelgpu
#include <omp.h>
diff --git a/offload/test/sanitizer/use_after_free_2.c b/offload/test/sanitizer/use_after_free_2.c
index 3ba621524e566..d47b8ef8c08e5 100644
--- a/offload/test/sanitizer/use_after_free_2.c
+++ b/offload/test/sanitizer/use_after_free_2.c
@@ -7,6 +7,7 @@
// UNSUPPORTED: nvidiagpu
//
// REQUIRES: gpu
+// https://github.com/llvm/llvm-project/issues/182119
// UNSUPPORTED: intelgpu
// If offload memory pooling is enabled for a large allocation, reuse error is
diff --git a/offload/test/unified_shared_memory/associate_ptr.c b/offload/test/unified_shared_memory/associate_ptr.c
index 08975ebef0025..402c6588ac547 100644
--- a/offload/test/unified_shared_memory/associate_ptr.c
+++ b/offload/test/unified_shared_memory/associate_ptr.c
@@ -2,7 +2,8 @@
// REQUIRES: unified_shared_memory
// UNSUPPORTED: clang-6, clang-7, clang-8, clang-9
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <assert.h>
#include <omp.h>
diff --git a/offload/test/unified_shared_memory/close_manual.c b/offload/test/unified_shared_memory/close_manual.c
index c588cb1c403a7..f1bd5791ed586 100644
--- a/offload/test/unified_shared_memory/close_manual.c
+++ b/offload/test/unified_shared_memory/close_manual.c
@@ -1,6 +1,8 @@
// RUN: %libomptarget-compile-run-and-check-generic
// REQUIRES: unified_shared_memory
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/unified_shared_memory/close_member.c b/offload/test/unified_shared_memory/close_member.c
index d4743a5872310..4f871eb3fa40f 100644
--- a/offload/test/unified_shared_memory/close_member.c
+++ b/offload/test/unified_shared_memory/close_member.c
@@ -2,7 +2,8 @@
// REQUIRES: unified_shared_memory
// UNSUPPORTED: clang-6, clang-7, clang-8, clang-9
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
>From ece290f83a567f5c780412f38ddcd9f73c0a00f7 Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Wed, 18 Mar 2026 11:54:54 -0700
Subject: [PATCH 11/17] Update failing tests
---
offload/test/api/assert.c | 3 ++-
offload/test/api/omp_indirect_call.c | 3 +--
offload/test/libc/malloc_parallel.c | 3 ++-
offload/test/mapping/lambda_by_value.cpp | 3 ++-
offload/test/mapping/present/zero_length_array_section_exit.c | 2 --
offload/test/mapping/use_device_addr/target_use_device_addr.c | 3 ++-
offload/test/offloading/assert.cpp | 3 ++-
offload/test/offloading/back2back_distribute.c | 3 ++-
offload/test/offloading/bug49334.cpp | 3 ++-
offload/test/offloading/bug51982.c | 3 ++-
offload/test/offloading/bug74582.c | 3 ++-
offload/test/offloading/ctor_dtor_lazy.cpp | 1 -
offload/test/offloading/dynamic_module.c | 2 ++
offload/test/offloading/malloc.c | 3 ++-
offload/test/offloading/spmdization.c | 3 ++-
offload/test/offloading/target_constexpr_mapping.cpp | 3 ++-
offload/test/sanitizer/ptr_outside_alloc_1.c | 3 ++-
17 files changed, 29 insertions(+), 18 deletions(-)
diff --git a/offload/test/api/assert.c b/offload/test/api/assert.c
index 9a9e770f19622..106a438b0dc82 100644
--- a/offload/test/api/assert.c
+++ b/offload/test/api/assert.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-run-and-check-generic
// RUN: %libomptarget-compileopt-run-and-check-generic
-
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <assert.h>
#include <stdio.h>
diff --git a/offload/test/api/omp_indirect_call.c b/offload/test/api/omp_indirect_call.c
index c74cdfc903066..b3e98d56a7706 100644
--- a/offload/test/api/omp_indirect_call.c
+++ b/offload/test/api/omp_indirect_call.c
@@ -1,6 +1,5 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
#include <assert.h>
#include <stdio.h>
diff --git a/offload/test/libc/malloc_parallel.c b/offload/test/libc/malloc_parallel.c
index 42f776826ba08..c74fc6720aadf 100644
--- a/offload/test/libc/malloc_parallel.c
+++ b/offload/test/libc/malloc_parallel.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-run-and-check-generic
// RUN: %libomptarget-compileopt-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/lambda_by_value.cpp b/offload/test/mapping/lambda_by_value.cpp
index b4e3f16c295a5..2978d66c17903 100644
--- a/offload/test/mapping/lambda_by_value.cpp
+++ b/offload/test/mapping/lambda_by_value.cpp
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compileopt-generic -fno-exceptions
// RUN: %libomptarget-run-generic 2>&1 | %fcheck-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdint.h>
#include <stdio.h>
diff --git a/offload/test/mapping/present/zero_length_array_section_exit.c b/offload/test/mapping/present/zero_length_array_section_exit.c
index 6d11e3accb6fb..5a7360542ec3c 100644
--- a/offload/test/mapping/present/zero_length_array_section_exit.c
+++ b/offload/test/mapping/present/zero_length_array_section_exit.c
@@ -1,8 +1,6 @@
// RUN: %libomptarget-compile-generic
// RUN: %libomptarget-run-fail-generic 2>&1 \
// RUN: | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/use_device_addr/target_use_device_addr.c b/offload/test/mapping/use_device_addr/target_use_device_addr.c
index 00a323f0499db..09bf5a63f3854 100644
--- a/offload/test/mapping/use_device_addr/target_use_device_addr.c
+++ b/offload/test/mapping/use_device_addr/target_use_device_addr.c
@@ -1,7 +1,8 @@
// RUN: %libomptarget-compile-generic -fopenmp-version=51
// RUN: %libomptarget-run-generic 2>&1 \
// RUN: | %fcheck-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
int main() {
diff --git a/offload/test/offloading/assert.cpp b/offload/test/offloading/assert.cpp
index 6ae33d1e6e6cf..20d0292d0e399 100644
--- a/offload/test/offloading/assert.cpp
+++ b/offload/test/offloading/assert.cpp
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compilexx-generic && %libomptarget-run-fail-generic
// RUN: %libomptarget-compileoptxx-generic && %libomptarget-run-fail-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
int main(int argc, char *argv[]) {
#pragma omp target
diff --git a/offload/test/offloading/back2back_distribute.c b/offload/test/offloading/back2back_distribute.c
index 4bd3d52a8c2fb..d5ab0500fe805 100644
--- a/offload/test/offloading/back2back_distribute.c
+++ b/offload/test/offloading/back2back_distribute.c
@@ -1,6 +1,7 @@
// clang-format off
// RUN: %libomptarget-compile-generic -O3 && %libomptarget-run-generic | %fcheck-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// clang-format on
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/bug49334.cpp b/offload/test/offloading/bug49334.cpp
index 47b79d408b93b..c7554b109c431 100644
--- a/offload/test/offloading/bug49334.cpp
+++ b/offload/test/offloading/bug49334.cpp
@@ -8,7 +8,8 @@
// REQUIRES: gpu
// UNSUPPORTED: nvidiagpu
// UNSUPPORTED: amdgpu
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <cassert>
#include <cmath>
diff --git a/offload/test/offloading/bug51982.c b/offload/test/offloading/bug51982.c
index a6910b1fbb3f8..ca32ff29437c9 100644
--- a/offload/test/offloading/bug51982.c
+++ b/offload/test/offloading/bug51982.c
@@ -1,7 +1,8 @@
// RUN: %libomptarget-compile-generic -O2 && %libomptarget-run-generic
// -O2 to run openmp-opt
// RUN: %libomptarget-compileopt-generic -O2 && %libomptarget-run-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
int main(void) {
long int aa = 0;
diff --git a/offload/test/offloading/bug74582.c b/offload/test/offloading/bug74582.c
index c6a283bb93691..a0fbfa261c57a 100644
--- a/offload/test/offloading/bug74582.c
+++ b/offload/test/offloading/bug74582.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-generic && %libomptarget-run-generic
// RUN: %libomptarget-compileopt-generic && %libomptarget-run-generic
-
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// Verify we do not read bits in the image that are not there (nobits section).
#pragma omp begin declare target
diff --git a/offload/test/offloading/ctor_dtor_lazy.cpp b/offload/test/offloading/ctor_dtor_lazy.cpp
index d09bd517387de..6ca232c288b5d 100644
--- a/offload/test/offloading/ctor_dtor_lazy.cpp
+++ b/offload/test/offloading/ctor_dtor_lazy.cpp
@@ -4,7 +4,6 @@
// RUN: %not --crash %libomptarget-run-generic
// RUN: %libomptarget-compilexx-generic -DCTOR_API
// RUN: %not --crash %libomptarget-run-generic
-// XFAIL: intelgpu
#include <cstdio>
#include <omp.h>
diff --git a/offload/test/offloading/dynamic_module.c b/offload/test/offloading/dynamic_module.c
index 3df81eb03648a..2da6b03b5a2f6 100644
--- a/offload/test/offloading/dynamic_module.c
+++ b/offload/test/offloading/dynamic_module.c
@@ -5,6 +5,8 @@
// RUN: %libomptarget-compileopt-generic %t.so && %libomptarget-run-generic 2>&1 | %fcheck-generic
//
// REQUIRES: gpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// clang-format on
diff --git a/offload/test/offloading/malloc.c b/offload/test/offloading/malloc.c
index d9b87d2bbe27c..fd912b8188f3a 100644
--- a/offload/test/offloading/malloc.c
+++ b/offload/test/offloading/malloc.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-generic && %libomptarget-run-generic
// RUN: %libomptarget-compileopt-generic && %libomptarget-run-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/offloading/spmdization.c b/offload/test/offloading/spmdization.c
index ca881a7580244..ea76d9c1a31c8 100644
--- a/offload/test/offloading/spmdization.c
+++ b/offload/test/offloading/spmdization.c
@@ -9,7 +9,8 @@
// clang-format on
// REQUIRES: gpu
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_constexpr_mapping.cpp b/offload/test/offloading/target_constexpr_mapping.cpp
index 240005b48763b..2175bf9072462 100644
--- a/offload/test/offloading/target_constexpr_mapping.cpp
+++ b/offload/test/offloading/target_constexpr_mapping.cpp
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compileoptxx-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/sanitizer/ptr_outside_alloc_1.c b/offload/test/sanitizer/ptr_outside_alloc_1.c
index d6e15b3aecb04..3e8ea09a05db9 100644
--- a/offload/test/sanitizer/ptr_outside_alloc_1.c
+++ b/offload/test/sanitizer/ptr_outside_alloc_1.c
@@ -9,7 +9,8 @@
// UNSUPPORTED: nvidiagpu
//
// REQUIRES: gpu
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
>From b4998f3dee8698eda9ef5e4f26e1c931976f6a1e Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Thu, 19 Mar 2026 13:46:08 -0700
Subject: [PATCH 12/17] Optimize implementation
---
openmp/device/src/Synchronization.cpp | 21 +++++++--------------
1 file changed, 7 insertions(+), 14 deletions(-)
diff --git a/openmp/device/src/Synchronization.cpp b/openmp/device/src/Synchronization.cpp
index 0fa6abd751f33..08522a96a1cf6 100644
--- a/openmp/device/src/Synchronization.cpp
+++ b/openmp/device/src/Synchronization.cpp
@@ -182,21 +182,21 @@ void setCriticalLock(omp_lock_t *Lock) { setLock(Lock); }
#if defined(__SPIRV__)
-[[clang::loader_uninitialized]] Local<uint32_t> namedBarrierTracker;
+[[clang::loader_uninitialized]] volatile Local<uint32_t> namedBarrierTracker;
void namedBarrierInit() {
- atomic::store(&namedBarrierTracker, 0u, atomic::seq_cst);
+ atomic::store(&namedBarrierTracker, 0u, atomic::seq_cst, atomic::workgroup);
}
void namedBarrier() {
uint32_t NumThreads = omp_get_num_threads();
- uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst);
+ uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst, atomic::workgroup);
- if (load == NumThreads - 1) {
- atomic::store(&namedBarrierTracker, 0, atomic::seq_cst);
+ if (load >= NumThreads - 1) {
+ atomic::store(&namedBarrierTracker, 0u, atomic::seq_cst, atomic::workgroup);
} else {
do {
- load = atomic::load(&namedBarrierTracker, atomic::seq_cst);
+ load = atomic::load(&namedBarrierTracker, atomic::seq_cst, atomic::workgroup);
} while (load != 0);
}
__gpu_sync_threads();
@@ -212,14 +212,7 @@ void initLock(omp_lock_t *Lock) { unsetLock(Lock); }
void destroyLock(omp_lock_t *Lock) { unsetLock(Lock); }
void setLock(omp_lock_t *Lock) {
int32_t *Lock_ptr = (int32_t *)Lock;
- bool Acquired = false;
- int32_t Expected;
- while (!Acquired) {
- Expected = 0;
- if (Expected == atomic::load(Lock_ptr, atomic::seq_cst))
- Acquired =
- atomic::cas(Lock_ptr, Expected, 1, atomic::seq_cst, atomic::seq_cst);
- }
+ while (!atomic::cas(Lock_ptr, 0, 1, atomic::seq_cst, atomic::seq_cst)) {}
}
void unsetCriticalLock(omp_lock_t *Lock) { unsetLock(Lock); }
>From d404db40e7df42d0bbc7ca070db8bf3c6df9aef1 Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Tue, 24 Mar 2026 11:17:37 -0700
Subject: [PATCH 13/17] Fix formatting
---
openmp/device/src/Synchronization.cpp | 9 ++++++---
1 file changed, 6 insertions(+), 3 deletions(-)
diff --git a/openmp/device/src/Synchronization.cpp b/openmp/device/src/Synchronization.cpp
index 08522a96a1cf6..c8168fbc357ee 100644
--- a/openmp/device/src/Synchronization.cpp
+++ b/openmp/device/src/Synchronization.cpp
@@ -190,13 +190,15 @@ void namedBarrierInit() {
void namedBarrier() {
uint32_t NumThreads = omp_get_num_threads();
- uint32_t load = atomic::add(&namedBarrierTracker, 1, atomic::seq_cst, atomic::workgroup);
+ uint32_t load =
+ atomic::add(&namedBarrierTracker, 1, atomic::seq_cst, atomic::workgroup);
if (load >= NumThreads - 1) {
atomic::store(&namedBarrierTracker, 0u, atomic::seq_cst, atomic::workgroup);
} else {
do {
- load = atomic::load(&namedBarrierTracker, atomic::seq_cst, atomic::workgroup);
+ load = atomic::load(&namedBarrierTracker, atomic::seq_cst,
+ atomic::workgroup);
} while (load != 0);
}
__gpu_sync_threads();
@@ -212,7 +214,8 @@ void initLock(omp_lock_t *Lock) { unsetLock(Lock); }
void destroyLock(omp_lock_t *Lock) { unsetLock(Lock); }
void setLock(omp_lock_t *Lock) {
int32_t *Lock_ptr = (int32_t *)Lock;
- while (!atomic::cas(Lock_ptr, 0, 1, atomic::seq_cst, atomic::seq_cst)) {}
+ while (!atomic::cas(Lock_ptr, 0, 1, atomic::seq_cst, atomic::seq_cst)) {
+ }
}
void unsetCriticalLock(omp_lock_t *Lock) { unsetLock(Lock); }
>From d2dcf8ae7174625317a711f8d7dfccf50f2ad11b Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Wed, 25 Mar 2026 17:16:52 -0700
Subject: [PATCH 14/17] Update tests
---
offload/test/api/assert.c | 2 --
offload/test/api/omp_device_uid.c | 2 --
offload/test/api/omp_get_device_num.c | 2 --
offload/test/api/omp_get_mapped_ptr.c | 2 --
offload/test/api/omp_get_num_devices.c | 2 --
offload/test/api/omp_get_num_procs.c | 2 --
offload/test/api/ompx_3d.c | 2 --
offload/test/api/ompx_3d.cpp | 2 --
offload/test/libc/malloc_parallel.c | 3 +--
offload/test/mapping/array_section_implicit_capture.c | 2 --
offload/test/mapping/chained_containing_structs_1.cc | 2 --
offload/test/mapping/chained_containing_structs_2.cc | 2 --
offload/test/mapping/chained_containing_structs_3.cc | 2 --
offload/test/mapping/data_member_ref.cpp | 2 --
.../test/mapping/declare_mapper_nested_default_mappers.cpp | 2 --
.../test/mapping/declare_mapper_nested_default_mappers_1.cpp | 4 +---
offload/test/mapping/declare_mapper_nested_mappers.cpp | 2 --
offload/test/mapping/declare_mapper_target_checks.cpp | 2 --
offload/test/mapping/delete_inf_refcount.c | 2 --
offload/test/mapping/duplicate_mappings_1.cpp | 2 --
offload/test/mapping/duplicate_mappings_2.cpp | 2 --
offload/test/mapping/firstprivate_aligned.cpp | 2 --
offload/test/mapping/implicit_device_ptr.c | 2 --
offload/test/mapping/lambda_by_value.cpp | 2 --
offload/test/mapping/low_alignment.c | 2 --
offload/test/mapping/map_both_pointer_pointee.c | 3 +--
offload/test/mapping/map_ordering_tgt_alloc_tofrom.c | 2 --
offload/test/mapping/map_ordering_tgt_data_alloc_from.c | 2 --
offload/test/mapping/map_ordering_tgt_data_alloc_to_from.c | 2 --
offload/test/mapping/map_ordering_tgt_data_alloc_tofrom.c | 2 --
offload/test/mapping/map_ordering_tgt_exit_data_delete_from.c | 2 --
offload/test/mapping/ompx_hold/struct.c | 2 --
offload/test/mapping/ompx_hold/target.c | 2 --
offload/test/mapping/pr38704.c | 2 --
offload/test/mapping/present/target.c | 2 --
offload/test/mapping/present/target_array_extension.c | 2 --
offload/test/mapping/present/unified_shared_memory.c | 2 --
offload/test/mapping/present/zero_length_array_section.c | 2 --
offload/test/mapping/ptr_and_obj_motion.c | 2 --
offload/test/mapping/target_data_array_extension_at_exit.c | 2 --
offload/test/mapping/target_derefence_array_pointrs.cpp | 3 +--
offload/test/mapping/target_has_device_addr.c | 3 +--
offload/test/mapping/target_implicit_partial_map.c | 2 --
offload/test/mapping/target_map_for_member_data.cpp | 2 --
offload/test/mapping/use_device_addr/target_use_device_addr.c | 2 --
offload/test/offloading/assert.cpp | 3 +--
offload/test/offloading/bug47654.cpp | 2 --
offload/test/offloading/bug49334.cpp | 3 +--
offload/test/offloading/bug50022.cpp | 2 --
offload/test/offloading/bug51781.c | 3 ++-
offload/test/offloading/bug53727.cpp | 2 --
offload/test/offloading/bug64959.c | 3 +--
offload/test/offloading/bug74582.c | 2 --
offload/test/offloading/ctor_dtor.cpp | 3 +--
offload/test/offloading/ctor_dtor_lazy.cpp | 1 +
offload/test/offloading/dynamic_module.c | 2 --
offload/test/offloading/dynamic_module_load.c | 2 --
offload/test/offloading/extern.c | 2 --
offload/test/offloading/force-usm.cpp | 3 +--
offload/test/offloading/generic_multiple_parallel_regions.c | 3 ++-
offload/test/offloading/global_constructor.cpp | 2 --
offload/test/offloading/host_as_target.c | 2 --
offload/test/offloading/info.c | 2 --
offload/test/offloading/malloc.c | 3 +--
offload/test/offloading/memory_manager.cpp | 2 --
offload/test/offloading/offloading_success.c | 2 --
offload/test/offloading/offloading_success.cpp | 2 --
offload/test/offloading/parallel_offloading_map.cpp | 2 --
offload/test/offloading/requires.c | 2 --
offload/test/offloading/static_linking.c | 2 --
offload/test/offloading/strided_multiple_update_from.c | 2 --
offload/test/offloading/strided_multiple_update_to.c | 2 --
offload/test/offloading/strided_partial_update_from.c | 2 --
offload/test/offloading/strided_partial_update_to.c | 2 --
offload/test/offloading/strided_ptr_multiple_update_from.c | 2 --
offload/test/offloading/strided_ptr_multiple_update_to.c | 2 --
offload/test/offloading/strided_ptr_partial_update_from.c | 2 --
offload/test/offloading/strided_ptr_partial_update_to.c | 2 --
offload/test/offloading/strided_update_count_expression.c | 2 --
.../test/offloading/strided_update_count_expression_complex.c | 3 ---
offload/test/offloading/strided_update_from.c | 2 --
.../strided_update_multiple_arrays_count_expression.c | 2 --
.../strided_update_multiple_arrays_variable_stride.c | 3 ---
offload/test/offloading/strided_update_to.c | 2 --
.../offloading/strided_update_variable_count_and_stride.c | 3 ---
offload/test/offloading/strided_update_variable_stride.c | 2 --
offload/test/offloading/strided_update_variable_stride_misc.c | 3 ---
offload/test/offloading/target-teams-atomic.c | 2 --
offload/test/offloading/target-tile.c | 2 --
offload/test/offloading/target_constexpr_mapping.cpp | 2 --
offload/test/offloading/target_critical_region.cpp | 3 +--
offload/test/offloading/target_depend_nowait.cpp | 2 --
offload/test/offloading/target_nowait_target.cpp | 2 --
offload/test/offloading/target_update_from.c | 2 --
offload/test/offloading/target_update_ptr_count_expression.c | 3 ---
.../offloading/target_update_ptr_variable_count_and_stride.c | 3 ---
offload/test/offloading/target_update_ptr_variable_stride.c | 3 ---
.../target_update_strided_struct_count_expression.c | 3 ---
offload/test/offloading/target_update_strided_struct_from.c | 2 --
.../offloading/target_update_strided_struct_multiple_from.c | 2 --
.../offloading/target_update_strided_struct_multiple_to.c | 2 --
.../offloading/target_update_strided_struct_partial_from.c | 2 --
.../test/offloading/target_update_strided_struct_partial_to.c | 2 --
offload/test/offloading/target_update_strided_struct_to.c | 2 --
.../target_update_strided_struct_variable_count_and_stride.c | 3 ---
.../offloading/target_update_strided_struct_variable_stride.c | 3 ---
offload/test/offloading/target_update_to.c | 2 --
offload/test/offloading/task_in_reduction_target.c | 2 --
offload/test/offloading/thread_state_1.c | 3 ++-
offload/test/offloading/thread_state_2.c | 3 ++-
offload/test/offloading/workshare_chunk.c | 3 +--
offload/test/sanitizer/ptr_outside_alloc_1.c | 3 +--
offload/test/sanitizer/use_after_free_1.c | 3 +--
offload/test/sanitizer/use_after_free_2.c | 3 +--
offload/test/unified_shared_memory/associate_ptr.c | 2 --
offload/test/unified_shared_memory/close_manual.c | 2 --
offload/test/unified_shared_memory/close_member.c | 2 --
117 files changed, 25 insertions(+), 239 deletions(-)
diff --git a/offload/test/api/assert.c b/offload/test/api/assert.c
index 106a438b0dc82..99d70810de0d6 100644
--- a/offload/test/api/assert.c
+++ b/offload/test/api/assert.c
@@ -1,7 +1,5 @@
// RUN: %libomptarget-compile-run-and-check-generic
// RUN: %libomptarget-compileopt-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <assert.h>
#include <stdio.h>
diff --git a/offload/test/api/omp_device_uid.c b/offload/test/api/omp_device_uid.c
index 0db41f76a73b2..0f9ad8e6009cd 100644
--- a/offload/test/api/omp_device_uid.c
+++ b/offload/test/api/omp_device_uid.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
#include <string.h>
diff --git a/offload/test/api/omp_get_device_num.c b/offload/test/api/omp_get_device_num.c
index b5c26cdf0d584..31e13b0050d41 100644
--- a/offload/test/api/omp_get_device_num.c
+++ b/offload/test/api/omp_get_device_num.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/api/omp_get_mapped_ptr.c b/offload/test/api/omp_get_mapped_ptr.c
index 9ba8fc3309edc..a8e11f912d668 100644
--- a/offload/test/api/omp_get_mapped_ptr.c
+++ b/offload/test/api/omp_get_mapped_ptr.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-and-run-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <assert.h>
#include <omp.h>
diff --git a/offload/test/api/omp_get_num_devices.c b/offload/test/api/omp_get_num_devices.c
index 97bd5b4927410..e8fbdcd1a1de0 100644
--- a/offload/test/api/omp_get_num_devices.c
+++ b/offload/test/api/omp_get_num_devices.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/api/omp_get_num_procs.c b/offload/test/api/omp_get_num_procs.c
index a0b9a0eaf4a4f..e958d0cf3492f 100644
--- a/offload/test/api/omp_get_num_procs.c
+++ b/offload/test/api/omp_get_num_procs.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/api/ompx_3d.c b/offload/test/api/ompx_3d.c
index 3dc902fe765d2..2ae17c226f3a4 100644
--- a/offload/test/api/ompx_3d.c
+++ b/offload/test/api/ompx_3d.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <ompx.h>
diff --git a/offload/test/api/ompx_3d.cpp b/offload/test/api/ompx_3d.cpp
index 75dec5fbf2f8e..5b5491263a917 100644
--- a/offload/test/api/ompx_3d.cpp
+++ b/offload/test/api/ompx_3d.cpp
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <ompx.h>
diff --git a/offload/test/libc/malloc_parallel.c b/offload/test/libc/malloc_parallel.c
index c74fc6720aadf..42f776826ba08 100644
--- a/offload/test/libc/malloc_parallel.c
+++ b/offload/test/libc/malloc_parallel.c
@@ -1,7 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
// RUN: %libomptarget-compileopt-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/array_section_implicit_capture.c b/offload/test/mapping/array_section_implicit_capture.c
index c2b80b201181a..1042bd23bd5ef 100644
--- a/offload/test/mapping/array_section_implicit_capture.c
+++ b/offload/test/mapping/array_section_implicit_capture.c
@@ -1,8 +1,6 @@
// RUN: %libomptarget-compile-generic
// RUN: %libomptarget-run-generic 2>&1 \
// RUN: | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/mapping/chained_containing_structs_1.cc b/offload/test/mapping/chained_containing_structs_1.cc
index dd4e26c1195ff..c9a09bb657be6 100644
--- a/offload/test/mapping/chained_containing_structs_1.cc
+++ b/offload/test/mapping/chained_containing_structs_1.cc
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <cstdlib>
#include <cstdio>
diff --git a/offload/test/mapping/chained_containing_structs_2.cc b/offload/test/mapping/chained_containing_structs_2.cc
index 247a75a8b569b..272b0ebaa7c92 100644
--- a/offload/test/mapping/chained_containing_structs_2.cc
+++ b/offload/test/mapping/chained_containing_structs_2.cc
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <cstdlib>
#include <cstdio>
diff --git a/offload/test/mapping/chained_containing_structs_3.cc b/offload/test/mapping/chained_containing_structs_3.cc
index 1c23b0e5a1755..763c9760918ef 100644
--- a/offload/test/mapping/chained_containing_structs_3.cc
+++ b/offload/test/mapping/chained_containing_structs_3.cc
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <cstdlib>
#include <cstdio>
diff --git a/offload/test/mapping/data_member_ref.cpp b/offload/test/mapping/data_member_ref.cpp
index b284706309adc..7947a62c169f4 100644
--- a/offload/test/mapping/data_member_ref.cpp
+++ b/offload/test/mapping/data_member_ref.cpp
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/declare_mapper_nested_default_mappers.cpp b/offload/test/mapping/declare_mapper_nested_default_mappers.cpp
index e9df27243a724..45fd042aedb0e 100644
--- a/offload/test/mapping/declare_mapper_nested_default_mappers.cpp
+++ b/offload/test/mapping/declare_mapper_nested_default_mappers.cpp
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <cstdio>
#include <cstdlib>
diff --git a/offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp b/offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp
index f337a593af828..a44e1b95fa99e 100644
--- a/offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp
+++ b/offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp
@@ -1,9 +1,7 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
// UNSUPPORTED: amdgcn-amd-amdhsa
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
-
+// XFAIL: intelgpu
extern "C" int printf(const char *, ...);
typedef struct {
diff --git a/offload/test/mapping/declare_mapper_nested_mappers.cpp b/offload/test/mapping/declare_mapper_nested_mappers.cpp
index 90c6d84ddab70..a59ed6980ec4c 100644
--- a/offload/test/mapping/declare_mapper_nested_mappers.cpp
+++ b/offload/test/mapping/declare_mapper_nested_mappers.cpp
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <cstdio>
#include <cstdlib>
diff --git a/offload/test/mapping/declare_mapper_target_checks.cpp b/offload/test/mapping/declare_mapper_target_checks.cpp
index 8f2f6c01419ab..019ad5dec61d3 100644
--- a/offload/test/mapping/declare_mapper_target_checks.cpp
+++ b/offload/test/mapping/declare_mapper_target_checks.cpp
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/delete_inf_refcount.c b/offload/test/mapping/delete_inf_refcount.c
index dd7bd7160e8ae..f6016c19b0ffb 100644
--- a/offload/test/mapping/delete_inf_refcount.c
+++ b/offload/test/mapping/delete_inf_refcount.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/duplicate_mappings_1.cpp b/offload/test/mapping/duplicate_mappings_1.cpp
index 9b77838d9ccdd..5bed3f161b4fe 100644
--- a/offload/test/mapping/duplicate_mappings_1.cpp
+++ b/offload/test/mapping/duplicate_mappings_1.cpp
@@ -1,7 +1,5 @@
// clang-format off
// RUN: %libomptarget-compilexx-generic -Wno-openmp-mapping && %libomptarget-run-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// clang-format on
diff --git a/offload/test/mapping/duplicate_mappings_2.cpp b/offload/test/mapping/duplicate_mappings_2.cpp
index 8122c48bd6be9..ca8112ee72544 100644
--- a/offload/test/mapping/duplicate_mappings_2.cpp
+++ b/offload/test/mapping/duplicate_mappings_2.cpp
@@ -1,7 +1,5 @@
// clang-format off
// RUN: %libomptarget-compilexx-generic -Wno-openmp-mapping && %libomptarget-run-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <assert.h>
diff --git a/offload/test/mapping/firstprivate_aligned.cpp b/offload/test/mapping/firstprivate_aligned.cpp
index 955e4ed708646..ae6be0f0c07f4 100644
--- a/offload/test/mapping/firstprivate_aligned.cpp
+++ b/offload/test/mapping/firstprivate_aligned.cpp
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-generic -O3 && %libomptarget-run-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/implicit_device_ptr.c b/offload/test/mapping/implicit_device_ptr.c
index 334265a1f8a82..baa75d21686a7 100644
--- a/offload/test/mapping/implicit_device_ptr.c
+++ b/offload/test/mapping/implicit_device_ptr.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/lambda_by_value.cpp b/offload/test/mapping/lambda_by_value.cpp
index 2978d66c17903..4c0278d405925 100644
--- a/offload/test/mapping/lambda_by_value.cpp
+++ b/offload/test/mapping/lambda_by_value.cpp
@@ -1,7 +1,5 @@
// RUN: %libomptarget-compileopt-generic -fno-exceptions
// RUN: %libomptarget-run-generic 2>&1 | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdint.h>
#include <stdio.h>
diff --git a/offload/test/mapping/low_alignment.c b/offload/test/mapping/low_alignment.c
index 3f2cceea1700a..615a5a9c31112 100644
--- a/offload/test/mapping/low_alignment.c
+++ b/offload/test/mapping/low_alignment.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/map_both_pointer_pointee.c b/offload/test/mapping/map_both_pointer_pointee.c
index 85326f6680d7b..0462beaf184df 100644
--- a/offload/test/mapping/map_both_pointer_pointee.c
+++ b/offload/test/mapping/map_both_pointer_pointee.c
@@ -3,8 +3,7 @@
// REQUIRES: unified_shared_memory
// UNSUPPORTED: amdgcn-amd-amdhsa
//
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
#pragma omp declare target
int *ptr1;
diff --git a/offload/test/mapping/map_ordering_tgt_alloc_tofrom.c b/offload/test/mapping/map_ordering_tgt_alloc_tofrom.c
index 3055722178ecc..806771e849167 100644
--- a/offload/test/mapping/map_ordering_tgt_alloc_tofrom.c
+++ b/offload/test/mapping/map_ordering_tgt_alloc_tofrom.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/map_ordering_tgt_data_alloc_from.c b/offload/test/mapping/map_ordering_tgt_data_alloc_from.c
index 2a98bcd0bd59f..b793551ee6ee4 100644
--- a/offload/test/mapping/map_ordering_tgt_data_alloc_from.c
+++ b/offload/test/mapping/map_ordering_tgt_data_alloc_from.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/map_ordering_tgt_data_alloc_to_from.c b/offload/test/mapping/map_ordering_tgt_data_alloc_to_from.c
index aa5a3b02535dd..c636cccdb9fa5 100644
--- a/offload/test/mapping/map_ordering_tgt_data_alloc_to_from.c
+++ b/offload/test/mapping/map_ordering_tgt_data_alloc_to_from.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/map_ordering_tgt_data_alloc_tofrom.c b/offload/test/mapping/map_ordering_tgt_data_alloc_tofrom.c
index f22303480ece1..0e44fc1531bdd 100644
--- a/offload/test/mapping/map_ordering_tgt_data_alloc_tofrom.c
+++ b/offload/test/mapping/map_ordering_tgt_data_alloc_tofrom.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/map_ordering_tgt_exit_data_delete_from.c b/offload/test/mapping/map_ordering_tgt_exit_data_delete_from.c
index 6455686a0f4a1..563cdfa3c3a61 100644
--- a/offload/test/mapping/map_ordering_tgt_exit_data_delete_from.c
+++ b/offload/test/mapping/map_ordering_tgt_exit_data_delete_from.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/ompx_hold/struct.c b/offload/test/mapping/ompx_hold/struct.c
index d46f4bbc68ccf..da2b38028762c 100644
--- a/offload/test/mapping/ompx_hold/struct.c
+++ b/offload/test/mapping/ompx_hold/struct.c
@@ -1,7 +1,5 @@
// RUN: %libomptarget-compile-generic -fopenmp-extensions
// RUN: %libomptarget-run-generic | %fcheck-generic -strict-whitespace
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/ompx_hold/target.c b/offload/test/mapping/ompx_hold/target.c
index e17e72cf07028..4b352cdb565d8 100644
--- a/offload/test/mapping/ompx_hold/target.c
+++ b/offload/test/mapping/ompx_hold/target.c
@@ -1,7 +1,5 @@
// RUN: %libomptarget-compile-generic -fopenmp-extensions
// RUN: %libomptarget-run-generic | %fcheck-generic -strict-whitespace
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/pr38704.c b/offload/test/mapping/pr38704.c
index 411eeaa12a6c7..72b4a171ccea4 100644
--- a/offload/test/mapping/pr38704.c
+++ b/offload/test/mapping/pr38704.c
@@ -3,8 +3,6 @@
// Clang 6.0 doesn't use the new map interface, undefined behavior when
// the compiler emits "old" interface code for structures.
// UNSUPPORTED: clang-6
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/mapping/present/target.c b/offload/test/mapping/present/target.c
index 63a9f9e0cdc3b..4344c42c808ff 100644
--- a/offload/test/mapping/present/target.c
+++ b/offload/test/mapping/present/target.c
@@ -1,8 +1,6 @@
// RUN: %libomptarget-compile-generic
// RUN: %libomptarget-run-fail-generic 2>&1 \
// RUN: | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/present/target_array_extension.c b/offload/test/mapping/present/target_array_extension.c
index 876c39aa0de01..873b2b36170ee 100644
--- a/offload/test/mapping/present/target_array_extension.c
+++ b/offload/test/mapping/present/target_array_extension.c
@@ -16,8 +16,6 @@
// RUN: %libomptarget-run-fail-generic 2>&1 \
// RUN: | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// END.
#include <stdio.h>
diff --git a/offload/test/mapping/present/unified_shared_memory.c b/offload/test/mapping/present/unified_shared_memory.c
index 06007615f1776..ab6e3bd0e5fa7 100644
--- a/offload/test/mapping/present/unified_shared_memory.c
+++ b/offload/test/mapping/present/unified_shared_memory.c
@@ -3,8 +3,6 @@
// RUN: | %fcheck-generic
// REQUIRES: unified_shared_memory
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/present/zero_length_array_section.c b/offload/test/mapping/present/zero_length_array_section.c
index 1a99617ed0165..e903a268ef335 100644
--- a/offload/test/mapping/present/zero_length_array_section.c
+++ b/offload/test/mapping/present/zero_length_array_section.c
@@ -1,8 +1,6 @@
// RUN: %libomptarget-compile-generic
// RUN: %libomptarget-run-fail-generic 2>&1 \
// RUN: | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/ptr_and_obj_motion.c b/offload/test/mapping/ptr_and_obj_motion.c
index d58c28a5fd4c2..a94c07aadc1bc 100644
--- a/offload/test/mapping/ptr_and_obj_motion.c
+++ b/offload/test/mapping/ptr_and_obj_motion.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/mapping/target_data_array_extension_at_exit.c b/offload/test/mapping/target_data_array_extension_at_exit.c
index 9cff8f8dabe0e..e300c800b2551 100644
--- a/offload/test/mapping/target_data_array_extension_at_exit.c
+++ b/offload/test/mapping/target_data_array_extension_at_exit.c
@@ -15,8 +15,6 @@
// RUN: -DEXTENDS=AFTER
// RUN: %libomptarget-run-generic 2>&1 \
// RUN: | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// END.
diff --git a/offload/test/mapping/target_derefence_array_pointrs.cpp b/offload/test/mapping/target_derefence_array_pointrs.cpp
index 4c464454a58cb..ed3c1ed17b243 100644
--- a/offload/test/mapping/target_derefence_array_pointrs.cpp
+++ b/offload/test/mapping/target_derefence_array_pointrs.cpp
@@ -6,8 +6,7 @@
// UNSUPPORTED: amdgcn-amd-amdhsa
// UNSUPPORTED: nvptx64-nvidia-cuda
// UNSUPPORTED: nvptx64-nvidia-cuda-LTO
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/mapping/target_has_device_addr.c b/offload/test/mapping/target_has_device_addr.c
index 21f373ad8dec6..5bff5168d4471 100644
--- a/offload/test/mapping/target_has_device_addr.c
+++ b/offload/test/mapping/target_has_device_addr.c
@@ -3,8 +3,7 @@
// RUN: | %fcheck-generic
// UNSUPPORTED: amdgcn-amd-amdhsa
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/target_implicit_partial_map.c b/offload/test/mapping/target_implicit_partial_map.c
index a9d40537916e1..ac04adf3d52dd 100644
--- a/offload/test/mapping/target_implicit_partial_map.c
+++ b/offload/test/mapping/target_implicit_partial_map.c
@@ -1,8 +1,6 @@
// RUN: %libomptarget-compile-generic
// RUN: %libomptarget-run-generic 2>&1 \
// RUN: | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// END.
#include <omp.h>
diff --git a/offload/test/mapping/target_map_for_member_data.cpp b/offload/test/mapping/target_map_for_member_data.cpp
index 5a31d546efd8f..fafff6b18a2bc 100644
--- a/offload/test/mapping/target_map_for_member_data.cpp
+++ b/offload/test/mapping/target_map_for_member_data.cpp
@@ -1,8 +1,6 @@
// RUN: %libomptarget-compile-generic -fopenmp-version=51
// RUN: %libomptarget-run-generic 2>&1 \
// RUN: | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
extern "C" int printf(const char *, ...);
template <typename T> class A {
diff --git a/offload/test/mapping/use_device_addr/target_use_device_addr.c b/offload/test/mapping/use_device_addr/target_use_device_addr.c
index 09bf5a63f3854..4a9dbe252f761 100644
--- a/offload/test/mapping/use_device_addr/target_use_device_addr.c
+++ b/offload/test/mapping/use_device_addr/target_use_device_addr.c
@@ -1,8 +1,6 @@
// RUN: %libomptarget-compile-generic -fopenmp-version=51
// RUN: %libomptarget-run-generic 2>&1 \
// RUN: | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
int main() {
diff --git a/offload/test/offloading/assert.cpp b/offload/test/offloading/assert.cpp
index 20d0292d0e399..6ae33d1e6e6cf 100644
--- a/offload/test/offloading/assert.cpp
+++ b/offload/test/offloading/assert.cpp
@@ -1,7 +1,6 @@
// RUN: %libomptarget-compilexx-generic && %libomptarget-run-fail-generic
// RUN: %libomptarget-compileoptxx-generic && %libomptarget-run-fail-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
int main(int argc, char *argv[]) {
#pragma omp target
diff --git a/offload/test/offloading/bug47654.cpp b/offload/test/offloading/bug47654.cpp
index 144321307536f..eba7904e76ca8 100644
--- a/offload/test/offloading/bug47654.cpp
+++ b/offload/test/offloading/bug47654.cpp
@@ -1,7 +1,5 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
// RUN: %libomptarget-compileoptxx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <cassert>
#include <iostream>
diff --git a/offload/test/offloading/bug49334.cpp b/offload/test/offloading/bug49334.cpp
index c7554b109c431..47b79d408b93b 100644
--- a/offload/test/offloading/bug49334.cpp
+++ b/offload/test/offloading/bug49334.cpp
@@ -8,8 +8,7 @@
// REQUIRES: gpu
// UNSUPPORTED: nvidiagpu
// UNSUPPORTED: amdgpu
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
#include <cassert>
#include <cmath>
diff --git a/offload/test/offloading/bug50022.cpp b/offload/test/offloading/bug50022.cpp
index ad6d9ece01782..d2ae02b7054b9 100644
--- a/offload/test/offloading/bug50022.cpp
+++ b/offload/test/offloading/bug50022.cpp
@@ -1,7 +1,5 @@
// RUN: %libomptarget-compilexx-and-run-generic
// RUN: %libomptarget-compileoptxx-and-run-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <cassert>
#include <iostream>
diff --git a/offload/test/offloading/bug51781.c b/offload/test/offloading/bug51781.c
index 3c7f2ce090224..b5841146788f3 100644
--- a/offload/test/offloading/bug51781.c
+++ b/offload/test/offloading/bug51781.c
@@ -33,7 +33,8 @@
// RUN: %fcheck-nvptx64-nvidia-cuda -check-prefix=CUSTOM -input-file=%t.custom
// RUN: %fcheck-amdgcn-amd-amdhsa -check-prefix=CUSTOM -input-file=%t.custom
// RUN: %libomptarget-run-generic 2>&1 | %fcheck-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
//
// CUSTOM: Rewriting generic-mode kernel with a customized state machine.
diff --git a/offload/test/offloading/bug53727.cpp b/offload/test/offloading/bug53727.cpp
index 9a81b530362e5..8ce8b7e9b87c4 100644
--- a/offload/test/offloading/bug53727.cpp
+++ b/offload/test/offloading/bug53727.cpp
@@ -1,7 +1,5 @@
// RUN: %libomptarget-compilexx-and-run-generic
// RUN: %libomptarget-compileoptxx-and-run-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <cassert>
#include <iostream>
diff --git a/offload/test/offloading/bug64959.c b/offload/test/offloading/bug64959.c
index 6b563f09087fd..a31b8449d34fc 100644
--- a/offload/test/offloading/bug64959.c
+++ b/offload/test/offloading/bug64959.c
@@ -6,8 +6,7 @@
// UNSUPPORTED: amdgcn-amd-amdhsa
// UNSUPPORTED: nvptx64-nvidia-cuda
// UNSUPPORTED: nvptx64-nvidia-cuda-LTO
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/bug74582.c b/offload/test/offloading/bug74582.c
index a0fbfa261c57a..7e85210dce092 100644
--- a/offload/test/offloading/bug74582.c
+++ b/offload/test/offloading/bug74582.c
@@ -1,7 +1,5 @@
// RUN: %libomptarget-compile-generic && %libomptarget-run-generic
// RUN: %libomptarget-compileopt-generic && %libomptarget-run-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// Verify we do not read bits in the image that are not there (nobits section).
#pragma omp begin declare target
diff --git a/offload/test/offloading/ctor_dtor.cpp b/offload/test/offloading/ctor_dtor.cpp
index 74561e4e1c516..d8ec93c7ae479 100644
--- a/offload/test/offloading/ctor_dtor.cpp
+++ b/offload/test/offloading/ctor_dtor.cpp
@@ -1,7 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
// RUN: %libomptarget-compileoptxx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
#include <cstdio>
struct S {
diff --git a/offload/test/offloading/ctor_dtor_lazy.cpp b/offload/test/offloading/ctor_dtor_lazy.cpp
index 6ca232c288b5d..d09bd517387de 100644
--- a/offload/test/offloading/ctor_dtor_lazy.cpp
+++ b/offload/test/offloading/ctor_dtor_lazy.cpp
@@ -4,6 +4,7 @@
// RUN: %not --crash %libomptarget-run-generic
// RUN: %libomptarget-compilexx-generic -DCTOR_API
// RUN: %not --crash %libomptarget-run-generic
+// XFAIL: intelgpu
#include <cstdio>
#include <omp.h>
diff --git a/offload/test/offloading/dynamic_module.c b/offload/test/offloading/dynamic_module.c
index 2da6b03b5a2f6..3df81eb03648a 100644
--- a/offload/test/offloading/dynamic_module.c
+++ b/offload/test/offloading/dynamic_module.c
@@ -5,8 +5,6 @@
// RUN: %libomptarget-compileopt-generic %t.so && %libomptarget-run-generic 2>&1 | %fcheck-generic
//
// REQUIRES: gpu
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// clang-format on
diff --git a/offload/test/offloading/dynamic_module_load.c b/offload/test/offloading/dynamic_module_load.c
index 9316c96bf5770..2e49cd35e5328 100644
--- a/offload/test/offloading/dynamic_module_load.c
+++ b/offload/test/offloading/dynamic_module_load.c
@@ -1,8 +1,6 @@
// clang-format off
// RUN: %libomptarget-compile-generic -DSHARED -fPIC -shared -o %t.so && %clang %flags %s -o %t -ldl && %libomptarget-run-generic %t.so 2>&1 | %fcheck-generic
// clang-format on
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#ifdef SHARED
#include <stdio.h>
int foo() {
diff --git a/offload/test/offloading/extern.c b/offload/test/offloading/extern.c
index 0a1af3367de78..4828855ee9c41 100644
--- a/offload/test/offloading/extern.c
+++ b/offload/test/offloading/extern.c
@@ -1,8 +1,6 @@
// clang-format off
// RUN: %libomptarget-compile-generic -DVAR -c -o %t.o
// RUN: %libomptarget-compile-generic %t.o && %libomptarget-run-generic | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// clang-format on
#ifdef VAR
int x = 1;
diff --git a/offload/test/offloading/force-usm.cpp b/offload/test/offloading/force-usm.cpp
index acb8a6d5ad070..9afe57d797356 100644
--- a/offload/test/offloading/force-usm.cpp
+++ b/offload/test/offloading/force-usm.cpp
@@ -10,8 +10,7 @@
//
// UNSUPPORTED: nvptx64-nvidia-cuda
// UNSUPPORTED: nvptx64-nvidia-cuda-LTO
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
// clang-format on
#include <cassert>
diff --git a/offload/test/offloading/generic_multiple_parallel_regions.c b/offload/test/offloading/generic_multiple_parallel_regions.c
index 80939d147ebca..cb7ea2d81b574 100644
--- a/offload/test/offloading/generic_multiple_parallel_regions.c
+++ b/offload/test/offloading/generic_multiple_parallel_regions.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-run-and-check-generic
// RUN: %libomptarget-compileopt-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/global_constructor.cpp b/offload/test/offloading/global_constructor.cpp
index 4892e50f572ca..9673e2b6b622b 100644
--- a/offload/test/offloading/global_constructor.cpp
+++ b/offload/test/offloading/global_constructor.cpp
@@ -1,7 +1,5 @@
// clang-format off
// RUN: %libomptarget-compilexx-generic && %libomptarget-run-generic | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// clang-format on
#include <cstdio>
diff --git a/offload/test/offloading/host_as_target.c b/offload/test/offloading/host_as_target.c
index a806bd181c01f..8ab991b723388 100644
--- a/offload/test/offloading/host_as_target.c
+++ b/offload/test/offloading/host_as_target.c
@@ -6,8 +6,6 @@
// - Works whether it's specified directly or as the default device.
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/info.c b/offload/test/offloading/info.c
index cebe99d093863..d86644b871e25 100644
--- a/offload/test/offloading/info.c
+++ b/offload/test/offloading/info.c
@@ -7,8 +7,6 @@
// FIXME: Fails due to optimized debugging in 'ptxas'.
// UNSUPPORTED: nvptx64-nvidia-cuda-LTO
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/malloc.c b/offload/test/offloading/malloc.c
index fd912b8188f3a..d9b87d2bbe27c 100644
--- a/offload/test/offloading/malloc.c
+++ b/offload/test/offloading/malloc.c
@@ -1,7 +1,6 @@
// RUN: %libomptarget-compile-generic && %libomptarget-run-generic
// RUN: %libomptarget-compileopt-generic && %libomptarget-run-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/offloading/memory_manager.cpp b/offload/test/offloading/memory_manager.cpp
index 811fe590507fd..d6d8697fcdec8 100644
--- a/offload/test/offloading/memory_manager.cpp
+++ b/offload/test/offloading/memory_manager.cpp
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
diff --git a/offload/test/offloading/offloading_success.c b/offload/test/offloading/offloading_success.c
index 89f33f6ac233f..f849cb4cfba79 100644
--- a/offload/test/offloading/offloading_success.c
+++ b/offload/test/offloading/offloading_success.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/offloading_success.cpp b/offload/test/offloading/offloading_success.cpp
index 41aefacaa8136..d3103b8666ab5 100644
--- a/offload/test/offloading/offloading_success.cpp
+++ b/offload/test/offloading/offloading_success.cpp
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/parallel_offloading_map.cpp b/offload/test/offloading/parallel_offloading_map.cpp
index fe61936508b1a..d2a6c4200df52 100644
--- a/offload/test/offloading/parallel_offloading_map.cpp
+++ b/offload/test/offloading/parallel_offloading_map.cpp
@@ -1,8 +1,6 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
// REQUIRES: gpu
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <cassert>
#include <iostream>
diff --git a/offload/test/offloading/requires.c b/offload/test/offloading/requires.c
index 188ed2d03f5ff..fc21ca3420e9e 100644
--- a/offload/test/offloading/requires.c
+++ b/offload/test/offloading/requires.c
@@ -1,8 +1,6 @@
// clang-format off
// RUN: %libomptarget-compile-generic -DREQ=1 && %libomptarget-run-generic 2>&1 | %fcheck-generic -check-prefix=GOOD
// RUN: %libomptarget-compile-generic -DREQ=2 && %not --crash %libomptarget-run-generic 2>&1 | %fcheck-generic -check-prefix=BAD
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// clang-format on
/*
diff --git a/offload/test/offloading/static_linking.c b/offload/test/offloading/static_linking.c
index d84ecbac74cd2..4487ab83dd5f1 100644
--- a/offload/test/offloading/static_linking.c
+++ b/offload/test/offloading/static_linking.c
@@ -2,8 +2,6 @@
// RUN: %libomptarget-compile-generic -DLIBRARY -c -o %t.o
// RUN: ar rcs %t.a %t.o
// RUN: %libomptarget-compile-generic %t.a && %libomptarget-run-generic 2>&1 | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// clang-format on
#ifdef LIBRARY
diff --git a/offload/test/offloading/strided_multiple_update_from.c b/offload/test/offloading/strided_multiple_update_from.c
index c048d0bcb21c4..a3e8d10863aef 100644
--- a/offload/test/offloading/strided_multiple_update_from.c
+++ b/offload/test/offloading/strided_multiple_update_from.c
@@ -3,8 +3,6 @@
// from the device to the host.
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_multiple_update_to.c b/offload/test/offloading/strided_multiple_update_to.c
index 1964cbf3f37cc..02b6210807d28 100644
--- a/offload/test/offloading/strided_multiple_update_to.c
+++ b/offload/test/offloading/strided_multiple_update_to.c
@@ -3,8 +3,6 @@
// from the host to the device.
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_partial_update_from.c b/offload/test/offloading/strided_partial_update_from.c
index 14ef6a882fee5..15d477f2b9b78 100644
--- a/offload/test/offloading/strided_partial_update_from.c
+++ b/offload/test/offloading/strided_partial_update_from.c
@@ -3,8 +3,6 @@
// across the array
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_partial_update_to.c b/offload/test/offloading/strided_partial_update_to.c
index bc53ab6f52d46..1cbdc0917c313 100644
--- a/offload/test/offloading/strided_partial_update_to.c
+++ b/offload/test/offloading/strided_partial_update_to.c
@@ -3,8 +3,6 @@
// across the array
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_ptr_multiple_update_from.c b/offload/test/offloading/strided_ptr_multiple_update_from.c
index e61b810cbc13d..b671a1e2b688d 100644
--- a/offload/test/offloading/strided_ptr_multiple_update_from.c
+++ b/offload/test/offloading/strided_ptr_multiple_update_from.c
@@ -3,8 +3,6 @@
// from the device to the host using dynamically allocated memory.
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/offloading/strided_ptr_multiple_update_to.c b/offload/test/offloading/strided_ptr_multiple_update_to.c
index 7fcf9c3f1267e..7ae9189ead8e3 100644
--- a/offload/test/offloading/strided_ptr_multiple_update_to.c
+++ b/offload/test/offloading/strided_ptr_multiple_update_to.c
@@ -3,8 +3,6 @@
// from the host to the device using dynamically allocated memory.
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/offloading/strided_ptr_partial_update_from.c b/offload/test/offloading/strided_ptr_partial_update_from.c
index c20f312a71b3c..d30a671393002 100644
--- a/offload/test/offloading/strided_ptr_partial_update_from.c
+++ b/offload/test/offloading/strided_ptr_partial_update_from.c
@@ -3,8 +3,6 @@
// across the array using dynamically allocated memory.
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/offloading/strided_ptr_partial_update_to.c b/offload/test/offloading/strided_ptr_partial_update_to.c
index 850f0f2e734a6..37c165351a760 100644
--- a/offload/test/offloading/strided_ptr_partial_update_to.c
+++ b/offload/test/offloading/strided_ptr_partial_update_to.c
@@ -3,8 +3,6 @@
// across the array using dynamically allocated memory.
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/offloading/strided_update_count_expression.c b/offload/test/offloading/strided_update_count_expression.c
index f005fde7dd4da..e18fc881a8021 100644
--- a/offload/test/offloading/strided_update_count_expression.c
+++ b/offload/test/offloading/strided_update_count_expression.c
@@ -4,8 +4,6 @@
// where the count (len/2) is a variable expression, not a constant.
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_count_expression_complex.c b/offload/test/offloading/strided_update_count_expression_complex.c
index 7f85dc1ac5203..54fdb342886f6 100644
--- a/offload/test/offloading/strided_update_count_expression_complex.c
+++ b/offload/test/offloading/strided_update_count_expression_complex.c
@@ -1,9 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
// Tests non-contiguous array sections with complex expression-based count
// scenarios including multiple struct arrays and non-zero offset.
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_from.c b/offload/test/offloading/strided_update_from.c
index 779834d6d37eb..fe875b7fd55c9 100644
--- a/offload/test/offloading/strided_update_from.c
+++ b/offload/test/offloading/strided_update_from.c
@@ -4,8 +4,6 @@
// other element (stride 2) from the device to the host
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_multiple_arrays_count_expression.c b/offload/test/offloading/strided_update_multiple_arrays_count_expression.c
index 8490b262dc47b..a1472cacc4a39 100644
--- a/offload/test/offloading/strided_update_multiple_arrays_count_expression.c
+++ b/offload/test/offloading/strided_update_multiple_arrays_count_expression.c
@@ -4,8 +4,6 @@
// same array with various count expressions.
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_multiple_arrays_variable_stride.c b/offload/test/offloading/strided_update_multiple_arrays_variable_stride.c
index 14662d8dcf7a7..0062d68cf2d57 100644
--- a/offload/test/offloading/strided_update_multiple_arrays_variable_stride.c
+++ b/offload/test/offloading/strided_update_multiple_arrays_variable_stride.c
@@ -1,9 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
// Tests multiple arrays with different variable strides in single update
// clause.
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_to.c b/offload/test/offloading/strided_update_to.c
index 0683f358f31f6..b90a240d1abe3 100644
--- a/offload/test/offloading/strided_update_to.c
+++ b/offload/test/offloading/strided_update_to.c
@@ -4,8 +4,6 @@
// other element (stride 2) from the host to the device
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_variable_count_and_stride.c b/offload/test/offloading/strided_update_variable_count_and_stride.c
index 09c1768bc1f8a..3600d7c94dff5 100644
--- a/offload/test/offloading/strided_update_variable_count_and_stride.c
+++ b/offload/test/offloading/strided_update_variable_count_and_stride.c
@@ -1,9 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
// Tests combining variable count expression AND variable stride in array
// sections.
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_variable_stride.c b/offload/test/offloading/strided_update_variable_stride.c
index 0cee4c832cefa..7c8079efa56ea 100644
--- a/offload/test/offloading/strided_update_variable_stride.c
+++ b/offload/test/offloading/strided_update_variable_stride.c
@@ -2,8 +2,6 @@
// Tests data[0:5:stride] where stride is a variable, making it non-contiguous.
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_variable_stride_misc.c b/offload/test/offloading/strided_update_variable_stride_misc.c
index fa0e754340985..df2bfa64cfe3c 100644
--- a/offload/test/offloading/strided_update_variable_stride_misc.c
+++ b/offload/test/offloading/strided_update_variable_stride_misc.c
@@ -1,9 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
// Miscellaneous variable stride tests: stride=1, stride=array_size, stride from
// array subscript.
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target-teams-atomic.c b/offload/test/offloading/target-teams-atomic.c
index f9b554438c2ba..97563422ced8f 100644
--- a/offload/test/offloading/target-teams-atomic.c
+++ b/offload/test/offloading/target-teams-atomic.c
@@ -3,8 +3,6 @@
// default.
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdbool.h>
diff --git a/offload/test/offloading/target-tile.c b/offload/test/offloading/target-tile.c
index e29da170cf940..8460b43b6f9c7 100644
--- a/offload/test/offloading/target-tile.c
+++ b/offload/test/offloading/target-tile.c
@@ -3,8 +3,6 @@
// RUN: %libomptarget-compile-generic -fopenmp-version=51
// RUN: %libomptarget-run-generic 2>&1 | %fcheck-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <stdio.h>
diff --git a/offload/test/offloading/target_constexpr_mapping.cpp b/offload/test/offloading/target_constexpr_mapping.cpp
index 2175bf9072462..14cf92a7cc26e 100644
--- a/offload/test/offloading/target_constexpr_mapping.cpp
+++ b/offload/test/offloading/target_constexpr_mapping.cpp
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compileoptxx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_critical_region.cpp b/offload/test/offloading/target_critical_region.cpp
index 836cce68e4fca..ddab2e754a66f 100644
--- a/offload/test/offloading/target_critical_region.cpp
+++ b/offload/test/offloading/target_critical_region.cpp
@@ -4,8 +4,7 @@
// UNSUPPORTED: nvptx64-nvidia-cuda
// UNSUPPORTED: nvptx64-nvidia-cuda-LTO
// UNSUPPORTED: amdgcn-amd-amdhsa
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_depend_nowait.cpp b/offload/test/offloading/target_depend_nowait.cpp
index fe060aac6ad95..b0d5408770bca 100644
--- a/offload/test/offloading/target_depend_nowait.cpp
+++ b/offload/test/offloading/target_depend_nowait.cpp
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compilexx-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_nowait_target.cpp b/offload/test/offloading/target_nowait_target.cpp
index a8cafaec46dbf..7c20e76ce138a 100644
--- a/offload/test/offloading/target_nowait_target.cpp
+++ b/offload/test/offloading/target_nowait_target.cpp
@@ -1,7 +1,5 @@
// RUN: %libomptarget-compilexx-and-run-generic
// RUN: %libomptarget-compileoptxx-and-run-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <cassert>
diff --git a/offload/test/offloading/target_update_from.c b/offload/test/offloading/target_update_from.c
index b10568a08f04d..4f84968f6d9a9 100644
--- a/offload/test/offloading/target_update_from.c
+++ b/offload/test/offloading/target_update_from.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// This test checks that "update from" clause in OpenMP supports strided
// sections. #pragma omp target update from(result[0:N/2:2]) updates every other
// element from device
diff --git a/offload/test/offloading/target_update_ptr_count_expression.c b/offload/test/offloading/target_update_ptr_count_expression.c
index 5c9af84f23983..c4b9fd566d401 100644
--- a/offload/test/offloading/target_update_ptr_count_expression.c
+++ b/offload/test/offloading/target_update_ptr_count_expression.c
@@ -1,9 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
// Tests non-contiguous array sections with expression-based count on
// heap-allocated pointer arrays with both FROM and TO directives.
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_update_ptr_variable_count_and_stride.c b/offload/test/offloading/target_update_ptr_variable_count_and_stride.c
index 5ab37159c4508..1a28595969c69 100644
--- a/offload/test/offloading/target_update_ptr_variable_count_and_stride.c
+++ b/offload/test/offloading/target_update_ptr_variable_count_and_stride.c
@@ -1,9 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
// Tests heap-allocated pointers with both variable count expression and
// variable stride.
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_update_ptr_variable_stride.c b/offload/test/offloading/target_update_ptr_variable_stride.c
index 04e657ea8bbe5..bea396065b760 100644
--- a/offload/test/offloading/target_update_ptr_variable_stride.c
+++ b/offload/test/offloading/target_update_ptr_variable_stride.c
@@ -1,9 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
// Tests non-contiguous array sections with variable stride on heap-allocated
// pointers.
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_update_strided_struct_count_expression.c b/offload/test/offloading/target_update_strided_struct_count_expression.c
index 2af7236ee688f..89b0e1b4c7ea7 100644
--- a/offload/test/offloading/target_update_strided_struct_count_expression.c
+++ b/offload/test/offloading/target_update_strided_struct_count_expression.c
@@ -1,9 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
// Tests non-contiguous array sections with expression-based count on struct
// member arrays with both FROM and TO directives.
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_update_strided_struct_from.c b/offload/test/offloading/target_update_strided_struct_from.c
index 9f13e61504a9a..46448d10189f2 100644
--- a/offload/test/offloading/target_update_strided_struct_from.c
+++ b/offload/test/offloading/target_update_strided_struct_from.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// This test checks that "update from" with user-defined mapper supports strided
// sections using fixed-size arrays in structs.
diff --git a/offload/test/offloading/target_update_strided_struct_multiple_from.c b/offload/test/offloading/target_update_strided_struct_multiple_from.c
index ae82ba04f9be0..fa57d35b86a4a 100644
--- a/offload/test/offloading/target_update_strided_struct_multiple_from.c
+++ b/offload/test/offloading/target_update_strided_struct_multiple_from.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// This test checks that #pragma omp target update from(s1.data[0:6:2],
// s2.data[0:4:3]) correctly updates strided sections covering the full arrays
// from device to host.
diff --git a/offload/test/offloading/target_update_strided_struct_multiple_to.c b/offload/test/offloading/target_update_strided_struct_multiple_to.c
index a61268f4e19e2..69c4015153580 100644
--- a/offload/test/offloading/target_update_strided_struct_multiple_to.c
+++ b/offload/test/offloading/target_update_strided_struct_multiple_to.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// This test checks that #pragma omp target update to(s1.data[0:6:2],
// s2.data[0:4:3]) correctly updates strided sections covering the full arrays
// from host to device.
diff --git a/offload/test/offloading/target_update_strided_struct_partial_from.c b/offload/test/offloading/target_update_strided_struct_partial_from.c
index 4f98a20f90373..c0ffc72f900a6 100644
--- a/offload/test/offloading/target_update_strided_struct_partial_from.c
+++ b/offload/test/offloading/target_update_strided_struct_partial_from.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// This test checks that #pragma omp target update from(s.data[0:2:3]) correctly
// updates every third element (stride 3) from the device to the host
// using a struct with fixed-size array member.
diff --git a/offload/test/offloading/target_update_strided_struct_partial_to.c b/offload/test/offloading/target_update_strided_struct_partial_to.c
index baff8a45b2dfd..018dcac7cdcf8 100644
--- a/offload/test/offloading/target_update_strided_struct_partial_to.c
+++ b/offload/test/offloading/target_update_strided_struct_partial_to.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// This test checks that #pragma omp target update to(s.data[0:2:3]) correctly
// updates every third element (stride 3) from the host to the device
// for struct member arrays.
diff --git a/offload/test/offloading/target_update_strided_struct_to.c b/offload/test/offloading/target_update_strided_struct_to.c
index b323740504fd1..90b3a5fe26989 100644
--- a/offload/test/offloading/target_update_strided_struct_to.c
+++ b/offload/test/offloading/target_update_strided_struct_to.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// This test checks that "update to" with struct member arrays supports strided
// sections using fixed-size arrays in structs.
diff --git a/offload/test/offloading/target_update_strided_struct_variable_count_and_stride.c b/offload/test/offloading/target_update_strided_struct_variable_count_and_stride.c
index 993f89ffa757c..6daf10383e921 100644
--- a/offload/test/offloading/target_update_strided_struct_variable_count_and_stride.c
+++ b/offload/test/offloading/target_update_strided_struct_variable_count_and_stride.c
@@ -1,9 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
// Tests struct member arrays with both variable count expression and variable
// stride.
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_update_strided_struct_variable_stride.c b/offload/test/offloading/target_update_strided_struct_variable_stride.c
index a2761b4d93470..4cd9da629ca93 100644
--- a/offload/test/offloading/target_update_strided_struct_variable_stride.c
+++ b/offload/test/offloading/target_update_strided_struct_variable_stride.c
@@ -1,9 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
// Tests non-contiguous array sections with variable stride on struct member
// arrays.
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/target_update_to.c b/offload/test/offloading/target_update_to.c
index 043b18018b951..d45d3182a166d 100644
--- a/offload/test/offloading/target_update_to.c
+++ b/offload/test/offloading/target_update_to.c
@@ -1,6 +1,4 @@
// RUN: %libomptarget-compile-run-and-check-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
// This test checks that "update to" clause in OpenMP supports strided sections.
// #pragma omp target update to(result[0:8:2]) updates every other element
// (stride 2)
diff --git a/offload/test/offloading/task_in_reduction_target.c b/offload/test/offloading/task_in_reduction_target.c
index 9281294128792..b546d73d544a5 100644
--- a/offload/test/offloading/task_in_reduction_target.c
+++ b/offload/test/offloading/task_in_reduction_target.c
@@ -1,7 +1,5 @@
// RUN: %libomptarget-compile-generic && \
// RUN: %libomptarget-run-generic
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/thread_state_1.c b/offload/test/offloading/thread_state_1.c
index 404ed6812928e..b5f17f0c45dca 100644
--- a/offload/test/offloading/thread_state_1.c
+++ b/offload/test/offloading/thread_state_1.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-run-and-check-generic
// RUN: %libomptarget-compileopt-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/thread_state_2.c b/offload/test/offloading/thread_state_2.c
index d40b987883c8f..e709f5f339447 100644
--- a/offload/test/offloading/thread_state_2.c
+++ b/offload/test/offloading/thread_state_2.c
@@ -1,6 +1,7 @@
// This fails when optimized for now.
// RUN: %libomptarget-compile-run-and-check-generic
-// XFAIL: intelgpu
+// https://github.com/llvm/llvm-project/issues/182119
+// UNSUPPORTED: intelgpu
// XUN: %libomptarget-compileopt-run-and-check-generic
#include <omp.h>
diff --git a/offload/test/offloading/workshare_chunk.c b/offload/test/offloading/workshare_chunk.c
index 4e8525282504a..551cc8d997cf7 100644
--- a/offload/test/offloading/workshare_chunk.c
+++ b/offload/test/offloading/workshare_chunk.c
@@ -2,8 +2,7 @@
// RUN: %libomptarget-compileopt-run-and-check-generic
// REQUIRES: gpu
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
// clang-format off
diff --git a/offload/test/sanitizer/ptr_outside_alloc_1.c b/offload/test/sanitizer/ptr_outside_alloc_1.c
index 3e8ea09a05db9..d6e15b3aecb04 100644
--- a/offload/test/sanitizer/ptr_outside_alloc_1.c
+++ b/offload/test/sanitizer/ptr_outside_alloc_1.c
@@ -9,8 +9,7 @@
// UNSUPPORTED: nvidiagpu
//
// REQUIRES: gpu
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
#include <omp.h>
diff --git a/offload/test/sanitizer/use_after_free_1.c b/offload/test/sanitizer/use_after_free_1.c
index 44a06f8506aa9..927e54ddab2bd 100644
--- a/offload/test/sanitizer/use_after_free_1.c
+++ b/offload/test/sanitizer/use_after_free_1.c
@@ -9,8 +9,7 @@
// UNSUPPORTED: nvidiagpu
//
// REQUIRES: gpu
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
#include <omp.h>
diff --git a/offload/test/sanitizer/use_after_free_2.c b/offload/test/sanitizer/use_after_free_2.c
index d47b8ef8c08e5..ece0cfe60875d 100644
--- a/offload/test/sanitizer/use_after_free_2.c
+++ b/offload/test/sanitizer/use_after_free_2.c
@@ -7,8 +7,7 @@
// UNSUPPORTED: nvidiagpu
//
// REQUIRES: gpu
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
+// XFAIL: intelgpu
// If offload memory pooling is enabled for a large allocation, reuse error is
// not detected. UNSUPPORTED: large_allocation_memory_pool
diff --git a/offload/test/unified_shared_memory/associate_ptr.c b/offload/test/unified_shared_memory/associate_ptr.c
index 402c6588ac547..0e9d25cd73546 100644
--- a/offload/test/unified_shared_memory/associate_ptr.c
+++ b/offload/test/unified_shared_memory/associate_ptr.c
@@ -2,8 +2,6 @@
// REQUIRES: unified_shared_memory
// UNSUPPORTED: clang-6, clang-7, clang-8, clang-9
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <assert.h>
#include <omp.h>
diff --git a/offload/test/unified_shared_memory/close_manual.c b/offload/test/unified_shared_memory/close_manual.c
index f1bd5791ed586..c588cb1c403a7 100644
--- a/offload/test/unified_shared_memory/close_manual.c
+++ b/offload/test/unified_shared_memory/close_manual.c
@@ -1,8 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
// REQUIRES: unified_shared_memory
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/unified_shared_memory/close_member.c b/offload/test/unified_shared_memory/close_member.c
index 4f871eb3fa40f..10ec42e90b02b 100644
--- a/offload/test/unified_shared_memory/close_member.c
+++ b/offload/test/unified_shared_memory/close_member.c
@@ -2,8 +2,6 @@
// REQUIRES: unified_shared_memory
// UNSUPPORTED: clang-6, clang-7, clang-8, clang-9
-// https://github.com/llvm/llvm-project/issues/182119
-// UNSUPPORTED: intelgpu
#include <omp.h>
#include <stdio.h>
>From 30401f0afd6b1e62a0ff1f2669a87d312e0b344a Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Wed, 25 Mar 2026 17:25:32 -0700
Subject: [PATCH 15/17] Remove spurious changes
---
offload/test/api/assert.c | 1 +
offload/test/api/omp_device_uid.c | 1 +
offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp | 1 +
offload/test/mapping/ompx_hold/target.c | 1 +
offload/test/mapping/target_implicit_partial_map.c | 1 +
offload/test/offloading/bug74582.c | 1 +
.../test/offloading/strided_update_count_expression_complex.c | 1 +
.../offloading/strided_update_multiple_arrays_variable_stride.c | 1 +
.../test/offloading/strided_update_variable_count_and_stride.c | 1 +
9 files changed, 9 insertions(+)
diff --git a/offload/test/api/assert.c b/offload/test/api/assert.c
index 99d70810de0d6..9a9e770f19622 100644
--- a/offload/test/api/assert.c
+++ b/offload/test/api/assert.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-run-and-check-generic
// RUN: %libomptarget-compileopt-run-and-check-generic
+
#include <assert.h>
#include <stdio.h>
diff --git a/offload/test/api/omp_device_uid.c b/offload/test/api/omp_device_uid.c
index 0f9ad8e6009cd..2a41d8d04ef8a 100644
--- a/offload/test/api/omp_device_uid.c
+++ b/offload/test/api/omp_device_uid.c
@@ -1,4 +1,5 @@
// RUN: %libomptarget-compile-run-and-check-generic
+
#include <omp.h>
#include <stdio.h>
#include <string.h>
diff --git a/offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp b/offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp
index a44e1b95fa99e..f822d08ba1e44 100644
--- a/offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp
+++ b/offload/test/mapping/declare_mapper_nested_default_mappers_1.cpp
@@ -2,6 +2,7 @@
// UNSUPPORTED: amdgcn-amd-amdhsa
// XFAIL: intelgpu
+
extern "C" int printf(const char *, ...);
typedef struct {
diff --git a/offload/test/mapping/ompx_hold/target.c b/offload/test/mapping/ompx_hold/target.c
index 4b352cdb565d8..40fd14c6bf005 100644
--- a/offload/test/mapping/ompx_hold/target.c
+++ b/offload/test/mapping/ompx_hold/target.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-generic -fopenmp-extensions
// RUN: %libomptarget-run-generic | %fcheck-generic -strict-whitespace
+
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/mapping/target_implicit_partial_map.c b/offload/test/mapping/target_implicit_partial_map.c
index ac04adf3d52dd..a8b2cc893df20 100644
--- a/offload/test/mapping/target_implicit_partial_map.c
+++ b/offload/test/mapping/target_implicit_partial_map.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-generic
// RUN: %libomptarget-run-generic 2>&1 \
// RUN: | %fcheck-generic
+
// END.
#include <omp.h>
diff --git a/offload/test/offloading/bug74582.c b/offload/test/offloading/bug74582.c
index 7e85210dce092..c6a283bb93691 100644
--- a/offload/test/offloading/bug74582.c
+++ b/offload/test/offloading/bug74582.c
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compile-generic && %libomptarget-run-generic
// RUN: %libomptarget-compileopt-generic && %libomptarget-run-generic
+
// Verify we do not read bits in the image that are not there (nobits section).
#pragma omp begin declare target
diff --git a/offload/test/offloading/strided_update_count_expression_complex.c b/offload/test/offloading/strided_update_count_expression_complex.c
index 54fdb342886f6..560c4e767f882 100644
--- a/offload/test/offloading/strided_update_count_expression_complex.c
+++ b/offload/test/offloading/strided_update_count_expression_complex.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Tests non-contiguous array sections with complex expression-based count
// scenarios including multiple struct arrays and non-zero offset.
+
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_multiple_arrays_variable_stride.c b/offload/test/offloading/strided_update_multiple_arrays_variable_stride.c
index 0062d68cf2d57..32859202a20fa 100644
--- a/offload/test/offloading/strided_update_multiple_arrays_variable_stride.c
+++ b/offload/test/offloading/strided_update_multiple_arrays_variable_stride.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Tests multiple arrays with different variable strides in single update
// clause.
+
#include <omp.h>
#include <stdio.h>
diff --git a/offload/test/offloading/strided_update_variable_count_and_stride.c b/offload/test/offloading/strided_update_variable_count_and_stride.c
index 3600d7c94dff5..1e1e41653c2c4 100644
--- a/offload/test/offloading/strided_update_variable_count_and_stride.c
+++ b/offload/test/offloading/strided_update_variable_count_and_stride.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-run-and-check-generic
// Tests combining variable count expression AND variable stride in array
// sections.
+
#include <omp.h>
#include <stdio.h>
>From 056e50c14c086b6b5802a5b45bf3e26fa01f8e2c Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Thu, 26 Mar 2026 14:13:32 -0700
Subject: [PATCH 16/17] Remove unnecessary volatile declaration
---
openmp/device/src/Synchronization.cpp | 2 +-
1 file changed, 1 insertion(+), 1 deletion(-)
diff --git a/openmp/device/src/Synchronization.cpp b/openmp/device/src/Synchronization.cpp
index c8168fbc357ee..244d42339918f 100644
--- a/openmp/device/src/Synchronization.cpp
+++ b/openmp/device/src/Synchronization.cpp
@@ -182,7 +182,7 @@ void setCriticalLock(omp_lock_t *Lock) { setLock(Lock); }
#if defined(__SPIRV__)
-[[clang::loader_uninitialized]] volatile Local<uint32_t> namedBarrierTracker;
+[[clang::loader_uninitialized]] Local<uint32_t> namedBarrierTracker;
void namedBarrierInit() {
atomic::store(&namedBarrierTracker, 0u, atomic::seq_cst, atomic::workgroup);
>From e3a601694caf9d93a5f5638fbb1646a0b0ee6d2d Mon Sep 17 00:00:00 2001
From: "Fine, Gregory" <gregory.fine at intel.com>
Date: Thu, 26 Mar 2026 16:37:18 -0700
Subject: [PATCH 17/17] Simplify implementation
---
openmp/device/src/Synchronization.cpp | 1 -
1 file changed, 1 deletion(-)
diff --git a/openmp/device/src/Synchronization.cpp b/openmp/device/src/Synchronization.cpp
index 244d42339918f..93f10ebd37292 100644
--- a/openmp/device/src/Synchronization.cpp
+++ b/openmp/device/src/Synchronization.cpp
@@ -201,7 +201,6 @@ void namedBarrier() {
atomic::workgroup);
} while (load != 0);
}
- __gpu_sync_threads();
}
void unsetLock(omp_lock_t *Lock) {
More information about the llvm-commits
mailing list