Skip to content

Commit 4f33d98

Browse files
Work-around to avoid failure from odr.cpp test
1 parent 7992db6 commit 4f33d98

File tree

3 files changed

+4
-3
lines changed

3 files changed

+4
-3
lines changed

sycl/include/sycl/ext/intel/esimd/detail/atomic_intrin.hpp

Lines changed: 2 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -205,7 +205,8 @@ template <typename Ty> Ty atomic_cmpxchg(Ty *ptr, Ty desired, Ty expected) {
205205
#endif
206206
}
207207

208-
void atomic_fence() {
208+
// Template argument is to avoid failure from 'odr.cpp' test
209+
template <int N> void atomic_fence() {
209210
#ifdef _WIN32
210211
// TODO: Windows will be supported soon
211212
__ESIMD_UNSUPPORTED_ON_HOST;

sycl/include/sycl/ext/intel/esimd/detail/memory_intrin.hpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -616,7 +616,7 @@ __ESIMD_INTRIN void __esimd_fence(uint8_t cntl)
616616
{
617617
// CM_EMU's 'cm_fence' is NOP. Disabled.
618618
// sycl::detail::getESIMDDeviceInterface()->cm_fence_ptr();
619-
__ESIMD_DNS::atomic_fence();
619+
__ESIMD_DNS::atomic_fence<0>();
620620
}
621621
#endif // __SYCL_DEVICE_ONLY__
622622

sycl/include/sycl/ext/intel/experimental/esimd/detail/memory_intrin.hpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1808,7 +1808,7 @@ __ESIMD_INTRIN void __esimd_lsc_fence(__ESIMD_DNS::simd_mask_storage_t<N> pred)
18081808
;
18091809
#else // __SYCL_DEVICE_ONLY__
18101810
{
1811-
__ESIMD_DNS::atomic_fence();
1811+
__ESIMD_DNS::atomic_fence<N>();
18121812
}
18131813
#endif // __SYCL_DEVICE_ONLY__
18141814

0 commit comments

Comments
 (0)