Skip to content

Commit 2cdcbed

Browse files
author
Pavel Samolysov
authored
[SYCL] Make sycl::ext::oneapi::atomic_ref available in sycl namespace (#4859)
* [SYCL] Add double as a valid atomic type In accordance to SYCL 2020 specification, section 4.15.3, the double data type is permitted to be a template parameter for atomic types. * [SYCL] Make sycl::ext::oneapi::atomic_ref available in sycl namespace The SYCL 2020 specification defines the sycl::atomic_ref template class (section 4.15.3). But the DPC++ compiler defines the template class under the sycl::ext::oneapi namespace. * [SYCL] Mark sycl::ext::oneapi::atomic_ref as deprecated * [SYCL] Extract IsValidAtomicRefType type trait Section 4.15.3. Atomic references specified valid atomic types for the parameter T of the sycl::atomic_ref template. * [SYCL] Add generic_space address space * [SYCL] Use generic_space address space as default in sycl::atomic_ref The SYCL 2020 version of atomic_ref defaults this last template argument (https://www.khronos.org/registry/SYCL/specs/sycl-2020/html/sycl-2020.html#sec:atomic-references): template <typename T, memory_order DefaultOrder, memory_scope DefaultScope, access::address_space Space = access::address_space::generic_space> class atomic_ref; * [SYCL] Trigger precommit tests * [SYCL] Return atomic_accessor back to use sycl::ext::oneapi::atomic_ref * [SYCL] Define sycl::ext::oneapi::atomic_ref as alias for sycl::atomic_ref * [SYCL] Remove memory_order constants from details of sycl::atomic_ref * Revert "[SYCL] Define sycl::ext::oneapi::atomic_ref as alias for sycl::atomic_ref" This reverts commit e44de23. * [SYCL] Add a test for deprecated sycl::ext::oneapi::atomic_ref * [SYCL] Fix a test for deprecated sycl::ext::oneapi::atomic_ref * [SYCL] Optimize imports in sycl::atomic_ref * [SYCL] Make memory_enums.hpp independent * [SYCL] Fix compilation errors when address_space::generic_space is used * [SYCL][NFC] Fix formatting issues * [SYCL] Fix a lot of llvm_unreachable("Invalid address space") errors Currently, CLang doesn't support the generic address space. Sema treats the generic address space as LangAS::Default: /// If this is an OpenCL address space attribute, returns its SYCL /// representation in LangAS, otherwise returns default address space. LangAS asSYCLLangAS() const { switch (getKind()) { ... case ParsedAttr::AT_OpenCLGenericAddressSpace: default: return LangAS::Default; } } and then this LangAS::Default leads to an invocation of llvm_unreachable: if (ASIdx == LangAS::Default) llvm_unreachable("Invalid address space"); As a temporal solution, I've defined `__OPENCL_GENERIC_AS__` as an empty string for host as well as for device. * [SYCL] Comment why we cannot use the opencl_generic attribute * [SYCL] Add opencl_generic as a valid address space for __SPIRV_ATOMICS * Revert "[SYCL] Add opencl_generic as a valid address space for __SPIRV_ATOMICS" This reverts commit 1dad8e9. * [SYCL] Optimize imports in sycl::ext::oneapi::atomic_ref * [SYCL] Emit an error when generic_space is used for sycl::atomic_ref The access::address_space::generic_space address space is not supported yet. * [SYCL] Configure the test to check generic_space not supporting * [SYCL] Fix the check generic_space not supporting * Revert "[SYCL] Comment why we cannot use the opencl_generic attribute" This reverts commit 0a41aa2. * Revert "[SYCL] Fix a lot of llvm_unreachable("Invalid address space") errors" This reverts commit 4402f47. * Revert "[SYCL][NFC] Fix formatting issues" This reverts commit ebe1f47. * Revert "[SYCL] Fix compilation errors when address_space::generic_space is used" This reverts commit e7f9906 * [SYCL][NFC] Comment that generic_space address space is not supported * [SYCL] Move the address space check into the atomic_ref_base class
1 parent 6dcc988 commit 2cdcbed

File tree

9 files changed

+802
-51
lines changed

9 files changed

+802
-51
lines changed

sycl/include/CL/sycl.hpp

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -12,6 +12,7 @@
1212
#include <CL/sycl/aspects.hpp>
1313
#include <CL/sycl/atomic.hpp>
1414
#include <CL/sycl/atomic_fence.hpp>
15+
#include <CL/sycl/atomic_ref.hpp>
1516
#include <CL/sycl/backend.hpp>
1617
#if SYCL_BACKEND_OPENCL
1718
#include <CL/sycl/backend/opencl.hpp>

sycl/include/CL/sycl/access/access.hpp

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -55,6 +55,7 @@ enum class address_space : int {
5555
global_host_space __SYCL2020_DEPRECATED(
5656
"use 'ext_intel_host_device_space' instead") =
5757
ext_intel_host_device_space,
58+
generic_space = 6, // TODO generic_space address space is not supported yet
5859
};
5960

6061
} // namespace access

0 commit comments

Comments
 (0)