[libc] [llvm] [OpenMP] Build OpenMP DeviceRTL on SPIRV64 (PR #121600)

[Nick Sarnie via llvm-commits]

https://github.com/sarnex created https://github.com/llvm/llvm-project/pull/121600

None

>From b2d153466d6ea1d29c1075eb27a1fae7ab14bfbe Mon Sep 17 00:00:00 2001
From: "Sarnie, Nick" <nick.sarnie at intel.com>
Date: Fri, 3 Jan 2025 12:31:00 -0800
Subject: [PATCH] [OpenMP] Build OpenMP DeviceRTL on SPIRV64

Signed-off-by: Sarnie, Nick <nick.sarnie at intel.com>
---
 libc/shared/rpc_util.h           | 2 +-
 offload/DeviceRTL/CMakeLists.txt | 3 +++
 2 files changed, 4 insertions(+), 1 deletion(-)

diff --git a/libc/shared/rpc_util.h b/libc/shared/rpc_util.h
index 9406de59f63b71..6b9df43776edac 100644
--- a/libc/shared/rpc_util.h
+++ b/libc/shared/rpc_util.h
@@ -157,7 +157,7 @@ RPC_ATTRS void sleep_briefly() {
     asm("nanosleep.u32 64;" ::: "memory");
 #elif defined(__AMDGPU__) && defined(RPC_TARGET_IS_GPU)
   __builtin_amdgcn_s_sleep(2);
-#elif __has_builtin(__builtin_ia32_pause)
+#elif __has_builtin(__builtin_ia32_pause) && !defined(__SPIRV__)
   __builtin_ia32_pause();
 #elif __has_builtin(__builtin_arm_isb)
   __builtin_arm_isb(0xf);
diff --git a/offload/DeviceRTL/CMakeLists.txt b/offload/DeviceRTL/CMakeLists.txt
index 22940264f9b19a..e707429a5196cf 100644
--- a/offload/DeviceRTL/CMakeLists.txt
+++ b/offload/DeviceRTL/CMakeLists.txt
@@ -264,6 +264,9 @@ compileDeviceRTLLibrary(amdgpu amdgcn-amd-amdhsa -Xclang -mcode-object-version=n
 add_custom_target(omptarget.devicertl.nvptx)
 compileDeviceRTLLibrary(nvptx nvptx64-nvidia-cuda --cuda-feature=+ptx63)
 
+add_custom_target(omptarget.devicertl.spirv64)
+compileDeviceRTLLibrary(spirv64 spirv64)
+
 # Archive all the object files generated above into a static library
 add_library(omptarget.devicertl STATIC)
 set_target_properties(omptarget.devicertl PROPERTIES