From 11186d6eadb35608f7017f83d72999d53f69b62b Mon Sep 17 00:00:00 2001 From: Andrei Elovikov Date: Fri, 2 Dec 2022 14:41:21 -0800 Subject: [PATCH 1/3] [NFC][SYCL] Switch to std::enable_if --- sycl/include/CL/__spirv/spirv_ops.hpp | 6 +- sycl/include/sycl/accessor.hpp | 208 ++++--- sycl/include/sycl/atomic.hpp | 8 +- sycl/include/sycl/atomic_ref.hpp | 6 +- sycl/include/sycl/backend/opencl.hpp | 14 +- sycl/include/sycl/buffer.hpp | 14 +- sycl/include/sycl/builtins.hpp | 560 +++++++++--------- sycl/include/sycl/detail/array.hpp | 10 +- sycl/include/sycl/detail/cg_types.hpp | 12 +- sycl/include/sycl/detail/common.hpp | 4 +- .../sycl/detail/generic_type_traits.hpp | 66 +-- sycl/include/sycl/detail/helpers.hpp | 10 +- .../sycl/detail/image_accessor_util.hpp | 18 +- .../sycl/detail/property_list_base.hpp | 24 +- sycl/include/sycl/detail/spirv.hpp | 83 +-- sycl/include/sycl/detail/stl_type_traits.hpp | 3 - .../sycl/detail/sycl_mem_obj_allocator.hpp | 4 +- sycl/include/sycl/detail/type_traits.hpp | 22 +- .../ext/intel/esimd/detail/memory_intrin.hpp | 2 +- .../ext/intel/esimd/detail/simd_obj_impl.hpp | 2 +- .../ext/intel/esimd/detail/simd_view_impl.hpp | 17 +- .../sycl/ext/intel/esimd/detail/sycl_util.hpp | 2 +- sycl/include/sycl/ext/intel/esimd/memory.hpp | 2 +- sycl/include/sycl/ext/intel/esimd/simd.hpp | 4 +- .../ext/intel/experimental/esimd/math.hpp | 4 +- .../ext/oneapi/accessor_property_list.hpp | 11 +- .../sycl/ext/oneapi/atomic_accessor.hpp | 6 +- sycl/include/sycl/ext/oneapi/atomic_ref.hpp | 6 +- .../sycl/ext/oneapi/backend/level_zero.hpp | 14 +- .../oneapi/device_global/device_global.hpp | 2 +- .../sycl/ext/oneapi/experimental/builtins.hpp | 8 +- .../ext/oneapi/experimental/spec_constant.hpp | 7 +- .../ext/oneapi/experimental/sycl_complex.hpp | 21 +- .../experimental/user_defined_reductions.hpp | 12 +- sycl/include/sycl/ext/oneapi/functional.hpp | 6 +- .../sycl/ext/oneapi/group_algorithm.hpp | 287 +++++---- .../sycl/ext/oneapi/sub_group_mask.hpp | 16 +- sycl/include/sycl/group.hpp | 40 +- sycl/include/sycl/group_algorithm.hpp | 212 +++---- sycl/include/sycl/handler.hpp | 46 +- sycl/include/sycl/id.hpp | 4 +- sycl/include/sycl/image.hpp | 18 +- sycl/include/sycl/item.hpp | 10 +- sycl/include/sycl/kernel_bundle.hpp | 4 +- sycl/include/sycl/multi_ptr.hpp | 86 +-- sycl/include/sycl/nd_item.hpp | 8 +- sycl/include/sycl/property_list.hpp | 2 +- sycl/include/sycl/range.hpp | 8 +- sycl/include/sycl/reduction.hpp | 84 +-- sycl/include/sycl/stream.hpp | 40 +- sycl/include/sycl/types.hpp | 195 +++--- sycl/source/detail/buffer_impl.hpp | 2 +- sycl/source/detail/builtins_geometric.cpp | 25 +- sycl/source/detail/builtins_relational.cpp | 7 +- .../source/detail/scheduler/graph_builder.cpp | 2 +- sycl/source/detail/scheduler/scheduler.hpp | 2 +- sycl/source/detail/sycl_mem_obj_t.hpp | 4 +- 57 files changed, 1118 insertions(+), 1182 deletions(-) diff --git a/sycl/include/CL/__spirv/spirv_ops.hpp b/sycl/include/CL/__spirv/spirv_ops.hpp index 7eaa886c74328..a25e4cec28c95 100644 --- a/sycl/include/CL/__spirv/spirv_ops.hpp +++ b/sycl/include/CL/__spirv/spirv_ops.hpp @@ -338,7 +338,7 @@ extern SYCL_EXTERNAL TempRetT __spirv_ImageSampleExplicitLod(SampledType, // of atomic min/max based on the type #define __SPIRV_ATOMIC_MINMAX(AS, Op) \ template \ - typename sycl::detail::enable_if_t< \ + typename std::enable_if_t< \ std::is_integral::value && std::is_signed::value, T> \ __spirv_Atomic##Op(AS T *Ptr, __spv::Scope::Flag Memory, \ __spv::MemorySemanticsMask::Flag Semantics, \ @@ -346,7 +346,7 @@ extern SYCL_EXTERNAL TempRetT __spirv_ImageSampleExplicitLod(SampledType, return __spirv_AtomicS##Op(Ptr, Memory, Semantics, Value); \ } \ template \ - typename sycl::detail::enable_if_t< \ + typename std::enable_if_t< \ std::is_integral::value && !std::is_signed::value, T> \ __spirv_Atomic##Op(AS T *Ptr, __spv::Scope::Flag Memory, \ __spv::MemorySemanticsMask::Flag Semantics, \ @@ -354,7 +354,7 @@ extern SYCL_EXTERNAL TempRetT __spirv_ImageSampleExplicitLod(SampledType, return __spirv_AtomicU##Op(Ptr, Memory, Semantics, Value); \ } \ template \ - typename sycl::detail::enable_if_t::value, T> \ + typename std::enable_if_t::value, T> \ __spirv_Atomic##Op(AS T *Ptr, __spv::Scope::Flag Memory, \ __spv::MemorySemanticsMask::Flag Semantics, \ T Value) { \ diff --git a/sycl/include/sycl/accessor.hpp b/sycl/include/sycl/accessor.hpp index 1874bc570ff17..4021615aa02d8 100644 --- a/sycl/include/sycl/accessor.hpp +++ b/sycl/include/sycl/accessor.hpp @@ -344,30 +344,28 @@ class accessor_common { MIDs[0] = Index; } - template 1)>> + template 1)>> auto operator[](size_t Index) { MIDs[Dims - CurDims] = Index; return AccessorSubscript(MAccessor, MIDs); } template > + typename = std::enable_if_t> RefType operator[](size_t Index) const { MIDs[Dims - CurDims] = Index; return MAccessor[MIDs]; } template - typename detail::enable_if_t> + typename std::enable_if_t> operator[](size_t Index) const { MIDs[Dims - CurDims] = Index; return MAccessor[MIDs]; } template > + typename = std::enable_if_t> ConstRefType operator[](size_t Index) const { MIDs[Dims - SubDims] = Index; return MAccessor[MIDs]; @@ -681,7 +679,7 @@ class image_accessor // accessor(image &imageRef); template < typename AllocatorT, int Dims = Dimensions, - typename = detail::enable_if_t<(Dims > 0 && Dims <= 3) && IsHostImageAcc>> + typename = std::enable_if_t<(Dims > 0 && Dims <= 3) && IsHostImageAcc>> image_accessor(image &ImageRef, int ImageElementSize) #ifdef __SYCL_DEVICE_ONLY__ { @@ -707,9 +705,8 @@ class image_accessor // template // accessor(image &imageRef, // handler &commandGroupHandlerRef); - template < - typename AllocatorT, int Dims = Dimensions, - typename = detail::enable_if_t<(Dims > 0 && Dims <= 3) && IsImageAcc>> + template 0 && Dims <= 3) && IsImageAcc>> image_accessor(image &ImageRef, handler &CommandGroupHandlerRef, int ImageElementSize) #ifdef __SYCL_DEVICE_ONLY__ @@ -763,17 +760,17 @@ class image_accessor size_t get_count() const { return size(); } size_t size() const noexcept { return get_range().size(); } - template > + template > range<1> get_range() const { cl_int Range = getRangeInternal(); return range<1>(Range); } - template > + template > range<2> get_range() const { cl_int2 Range = getRangeInternal(); return range<2>(Range[0], Range[1]); } - template > + template > range<3> get_range() const { cl_int3 Range = getRangeInternal(); return range<3>(Range[0], Range[1], Range[2]); @@ -784,7 +781,7 @@ class image_accessor size_t get_count() const { return size(); }; size_t size() const noexcept { return MImageCount; }; - template 0)>> + template 0)>> range get_range() const { return detail::convertToArrayOfN(getAccessRange()); } @@ -796,7 +793,7 @@ class image_accessor // || (accessTarget == access::target::host_image && ( accessMode == // access::mode::read || accessMode == access::mode::read_write)) template 0) && (IsValidCoordDataT::value) && (detail::is_genint::value) && ((IsImageAcc && IsImageAccessReadOnly) || @@ -816,7 +813,7 @@ class image_accessor // || (accessTarget == access::target::host_image && ( accessMode == // access::mode::read || accessMode == access::mode::read_write)) template 0) && (IsValidCoordDataT::value) && ((IsImageAcc && IsImageAccessReadOnly) || (IsHostImageAcc && IsImageAccessAnyRead))>> @@ -840,7 +837,7 @@ class image_accessor // access::mode::write || accessMode == access::mode::discard_write || // accessMode == access::mode::read_write)) template 0) && (detail::is_genint::value) && (IsValidCoordDataT::value) && ((IsImageAcc && IsImageAccessWriteOnly) || @@ -896,22 +893,22 @@ class __image_array_slice__ { : MBaseAcc(BaseAcc), MIdx(Idx) {} template 0) && (IsValidCoordDataT::value)>> DataT read(const CoordT &Coords) const { return MBaseAcc.read(getAdjustedCoords(Coords)); } template 0) && IsValidCoordDataT::value>> + typename = std::enable_if_t<(Dims > 0) && + IsValidCoordDataT::value>> DataT read(const CoordT &Coords, const sampler &Smpl) const { return MBaseAcc.read(getAdjustedCoords(Coords), Smpl); } template 0) && IsValidCoordDataT::value>> + typename = std::enable_if_t<(Dims > 0) && + IsValidCoordDataT::value>> void write(const CoordT &Coords, const DataT &Color) const { return MBaseAcc.write(getAdjustedCoords(Coords), Color); } @@ -921,12 +918,12 @@ class __image_array_slice__ { size_t get_count() const { return size(); } size_t size() const noexcept { return get_range().size(); } - template > + template > range<1> get_range() const { cl_int2 Count = MBaseAcc.getRangeInternal(); return range<1>(Count.x()); } - template > + template > range<2> get_range() const { cl_int3 Count = MBaseAcc.getRangeInternal(); return range<2>(Count.x(), Count.y()); @@ -941,7 +938,7 @@ class __image_array_slice__ { } template > + typename = std::enable_if_t<(Dims == 1 || Dims == 2)>> range get_range() const { return detail::convertToArrayOfN(MBaseAcc.getAccessRange()); } @@ -1244,7 +1241,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : public: template ::value && std::is_same::value && Dims == 0 && ((!IsPlaceH && IsHostBuf) || @@ -1276,7 +1273,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && std::is_same::value && Dims == 0 && ((!IsPlaceH && IsHostBuf) || @@ -1308,7 +1305,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : } template ::value && std::is_same::value && (Dims == 0) && (!IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf))>> @@ -1340,7 +1337,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && std::is_same::value && (Dims == 0) && (!IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf))>> @@ -1372,7 +1369,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : #endif template ::value && IsSameAsBuffer() && ((!IsPlaceH && IsHostBuf) || @@ -1406,7 +1403,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && ((!IsPlaceH && IsHostBuf) || @@ -1443,7 +1440,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && IsValidTag() && IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf)>> @@ -1457,7 +1454,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && IsValidTag() && IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf)>> @@ -1472,7 +1469,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : #endif template ::value && IsSameAsBuffer() && (!IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf))>> @@ -1505,7 +1502,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && (!IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf))>> @@ -1541,7 +1538,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && IsValidTag() && !IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf)>> @@ -1555,7 +1552,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && IsValidTag() && !IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf)>> @@ -1572,7 +1569,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : #endif template ::value && IsSameAsBuffer() && ((!IsPlaceH && IsHostBuf) || @@ -1585,7 +1582,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && ((!IsPlaceH && IsHostBuf) || @@ -1601,7 +1598,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && IsValidTag() && IsPlaceH && (IsGlobalBuf || IsConstantBuf)>> @@ -1615,7 +1612,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && IsValidTag() && IsPlaceH && (IsGlobalBuf || IsConstantBuf)>> @@ -1631,7 +1628,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : #endif template ::value && IsSameAsBuffer() && (!IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf))>> @@ -1644,7 +1641,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && (!IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf))>> @@ -1661,7 +1658,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && IsValidTag() && !IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf)>> @@ -1677,7 +1674,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && IsValidTag() && !IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf)>> @@ -1694,7 +1691,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : #endif template ::value && IsSameAsBuffer() && ((!IsPlaceH && IsHostBuf) || @@ -1735,7 +1732,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && ((!IsPlaceH && IsHostBuf) || @@ -1780,7 +1777,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && IsValidTag() && IsPlaceH && (IsGlobalBuf || IsConstantBuf)>> @@ -1794,7 +1791,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && IsValidTag() && IsPlaceH && (IsGlobalBuf || IsConstantBuf)>> @@ -1810,7 +1807,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : #endif template ::value && IsSameAsBuffer() && (!IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf))>> @@ -1851,7 +1848,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && (!IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf))>> @@ -1895,7 +1892,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && IsValidTag() && !IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf)>> @@ -1911,7 +1908,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : template ::value && IsSameAsBuffer() && IsValidTag() && !IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf)>> @@ -1972,12 +1969,12 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : bool empty() const noexcept { return size() == 0; } - template 0)>> + template 0)>> range get_range() const { return detail::convertToArrayOfN(getAccessRange()); } - template 0)>> + template 0)>> id get_offset() const { #if __cplusplus >= 201703L static_assert( @@ -1989,51 +1986,51 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : } template ::value>> + typename = std::enable_if_t::value>> operator RefType() const { const size_t LinearIndex = getLinearIndex(id()); return *(getQualifiedPtr() + LinearIndex); } template > + typename = std::enable_if_t> operator ConstRefType() const { const size_t LinearIndex = getLinearIndex(id()); return *(getQualifiedPtr() + LinearIndex); } template 0) && IsAccessAnyWrite>> + typename = std::enable_if_t<(Dims > 0) && IsAccessAnyWrite>> RefType operator[](id Index) const { const size_t LinearIndex = getLinearIndex(Index); return getQualifiedPtr()[LinearIndex]; } template - typename detail::enable_if_t<(Dims > 0) && IsAccessReadOnly, ConstRefType> + typename std::enable_if_t<(Dims > 0) && IsAccessReadOnly, ConstRefType> operator[](id Index) const { const size_t LinearIndex = getLinearIndex(Index); return getQualifiedPtr()[LinearIndex]; } template - operator typename detail::enable_if_t + atomic #else - atomic + atomic #endif - >() const { + >() const { const size_t LinearIndex = getLinearIndex(id()); return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } template - typename detail::enable_if_t<(Dims > 0) && AccessMode == access::mode::atomic, - atomic> + typename std::enable_if_t<(Dims > 0) && AccessMode == access::mode::atomic, + atomic> operator[](id Index) const { const size_t LinearIndex = getLinearIndex(Index); return atomic(multi_ptr( @@ -2041,35 +2038,35 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : } template - typename detail::enable_if_t> + typename std::enable_if_t> operator[](size_t Index) const { const size_t LinearIndex = getLinearIndex(id(Index)); return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } - template 1)>> + template 1)>> auto operator[](size_t Index) const { return AccessorSubscript(*this, Index); } - template > + template < + access::target AccessTarget_ = AccessTarget, + typename = std::enable_if_t> DataT *get_pointer() const { return getPointerAdjusted(); } template < access::target AccessTarget_ = AccessTarget, - typename = detail::enable_if_t> + typename = std::enable_if_t> global_ptr get_pointer() const { return global_ptr(getPointerAdjusted()); } template > + typename = std::enable_if_t> constant_ptr get_pointer() const { return constant_ptr(getPointerAdjusted()); } @@ -2078,7 +2075,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : // code. This restriction is not listed in the core spec and will be added in // future versions. template - typename sycl::detail::enable_if_t< + typename std::enable_if_t< !ext::oneapi::is_compile_time_property::value, bool> has_property() const noexcept { #ifndef __SYCL_DEVICE_ONLY__ @@ -2092,7 +2089,7 @@ class __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : // code. This restriction is not listed in the core spec and will be added in // future versions. template ::value>> Property get_property() const { #ifndef __SYCL_DEVICE_ONLY__ @@ -2476,7 +2473,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : using reference = DataT &; using const_reference = const DataT &; - template > + template > local_accessor_base(handler &, const detail::code_location CodeLoc = detail::code_location::current()) #ifdef __SYCL_DEVICE_ONLY__ @@ -2489,8 +2486,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : } #endif - template > + template > local_accessor_base(handler &, const property_list &propList, const detail::code_location CodeLoc = detail::code_location::current()) @@ -2507,7 +2503,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : } #endif - template 0)>> + template 0)>> local_accessor_base( range AllocationSize, handler &, const detail::code_location CodeLoc = detail::code_location::current()) @@ -2523,7 +2519,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : #endif template 0)>> + typename = std::enable_if_t<(Dims > 0)>> local_accessor_base(range AllocationSize, handler &, const property_list &propList, const detail::code_location CodeLoc = @@ -2547,32 +2543,32 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : size_t get_count() const { return size(); } size_t size() const noexcept { return getSize().size(); } - template 0)>> + template 0)>> range get_range() const { return detail::convertToArrayOfN(getSize()); } template > + typename = std::enable_if_t> operator RefType() const { return *getQualifiedPtr(); } template 0) && IsAccessAnyWrite>> + typename = std::enable_if_t<(Dims > 0) && IsAccessAnyWrite>> RefType operator[](id Index) const { const size_t LinearIndex = getLinearIndex(Index); return getQualifiedPtr()[LinearIndex]; } template > + typename = std::enable_if_t> RefType operator[](size_t Index) const { return getQualifiedPtr()[Index]; } template - operator typename detail::enable_if_t< + operator typename std::enable_if_t< Dims == 0 && AccessMode == access::mode::atomic, atomic>() const { return atomic( @@ -2580,8 +2576,8 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : } template - typename detail::enable_if_t<(Dims > 0) && AccessMode == access::mode::atomic, - atomic> + typename std::enable_if_t<(Dims > 0) && AccessMode == access::mode::atomic, + atomic> operator[](id Index) const { const size_t LinearIndex = getLinearIndex(Index); return atomic(multi_ptr( @@ -2589,14 +2585,14 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : } template - typename detail::enable_if_t> + typename std::enable_if_t> operator[](size_t Index) const { return atomic(multi_ptr( getQualifiedPtr() + Index)); } - template 1)>> + template 1)>> typename AccessorCommonT::template AccessorSubscript< Dims - 1, local_accessor_base> @@ -2945,7 +2941,7 @@ class host_accessor // -------+---------+-------+----+----------+-------------- template ::value && Dims == 0>> host_accessor( buffer &BufferRef, @@ -2954,7 +2950,7 @@ class host_accessor : AccessorT(BufferRef, PropertyList, CodeLoc) {} template ()>> + typename = std::enable_if_t()>> host_accessor( buffer &BufferRef, const property_list &PropertyList = {}, @@ -2964,7 +2960,7 @@ class host_accessor #if __cplusplus >= 201703L template ()>> + typename = std::enable_if_t()>> host_accessor( buffer &BufferRef, mode_tag_t, const property_list &PropertyList = {}, @@ -2974,7 +2970,7 @@ class host_accessor #endif template ()>> + typename = std::enable_if_t()>> host_accessor( buffer &BufferRef, handler &CommandGroupHandler, const property_list &PropertyList = {}, @@ -2984,7 +2980,7 @@ class host_accessor #if __cplusplus >= 201703L template ()>> + typename = std::enable_if_t()>> host_accessor( buffer &BufferRef, handler &CommandGroupHandler, mode_tag_t, @@ -2995,7 +2991,7 @@ class host_accessor #endif template ()>> + typename = std::enable_if_t()>> host_accessor( buffer &BufferRef, range AccessRange, const property_list &PropertyList = {}, @@ -3005,7 +3001,7 @@ class host_accessor #if __cplusplus >= 201703L template ()>> + typename = std::enable_if_t()>> host_accessor( buffer &BufferRef, range AccessRange, mode_tag_t, @@ -3016,7 +3012,7 @@ class host_accessor #endif template ()>> + typename = std::enable_if_t()>> host_accessor( buffer &BufferRef, handler &CommandGroupHandler, range AccessRange, @@ -3028,7 +3024,7 @@ class host_accessor #if __cplusplus >= 201703L template ()>> + typename = std::enable_if_t()>> host_accessor( buffer &BufferRef, handler &CommandGroupHandler, range AccessRange, @@ -3040,7 +3036,7 @@ class host_accessor #endif template ()>> + typename = std::enable_if_t()>> host_accessor( buffer &BufferRef, range AccessRange, id AccessOffset, @@ -3052,7 +3048,7 @@ class host_accessor #if __cplusplus >= 201703L template ()>> + typename = std::enable_if_t()>> host_accessor( buffer &BufferRef, range AccessRange, id AccessOffset, @@ -3064,7 +3060,7 @@ class host_accessor #endif template ()>> + typename = std::enable_if_t()>> host_accessor( buffer &BufferRef, handler &CommandGroupHandler, range AccessRange, @@ -3076,7 +3072,7 @@ class host_accessor #if __cplusplus >= 201703L template ()>> + typename = std::enable_if_t()>> host_accessor( buffer &BufferRef, handler &CommandGroupHandler, range AccessRange, diff --git a/sycl/include/sycl/atomic.hpp b/sycl/include/sycl/atomic.hpp index 697907dc0c45b..beb9174941e30 100644 --- a/sycl/include/sycl/atomic.hpp +++ b/sycl/include/sycl/atomic.hpp @@ -210,7 +210,7 @@ class __SYCL2020_DEPRECATED( #ifdef __ENABLE_USM_ADDR_SPACE__ // Create atomic in global_space with one from ext_intel_global_device_space template > atomic(const atomic @@ -219,7 +219,7 @@ class __SYCL2020_DEPRECATED( } template > atomic( @@ -235,13 +235,13 @@ class __SYCL2020_DEPRECATED( #ifdef __SYCL_DEVICE_ONLY__ template - detail::enable_if_t::value, T> + std::enable_if_t::value, T> load(memory_order Order = memory_order::relaxed) const { return __spirv_AtomicLoad(Ptr, SpirvScope, detail::getSPIRVMemorySemanticsMask(Order)); } template - detail::enable_if_t::value, T> + std::enable_if_t::value, T> load(memory_order Order = memory_order::relaxed) const { auto *TmpPtr = reinterpret_cast::pointer>(Ptr); diff --git a/sycl/include/sycl/atomic_ref.hpp b/sycl/include/sycl/atomic_ref.hpp index cbae52de9d7c7..f20549f847e58 100644 --- a/sycl/include/sycl/atomic_ref.hpp +++ b/sycl/include/sycl/atomic_ref.hpp @@ -89,7 +89,7 @@ inline constexpr memory_order getLoadOrder(memory_order order) { template struct bit_equal; template -struct bit_equal::value>> { +struct bit_equal::value>> { bool operator()(const T &lhs, const T &rhs) { return lhs == rhs; } }; @@ -266,7 +266,7 @@ class atomic_ref_impl template class atomic_ref_impl::value>> + typename std::enable_if_t::value>> : public atomic_ref_base { public: @@ -414,7 +414,7 @@ template class atomic_ref_impl< T, DefaultOrder, DefaultScope, AddressSpace, - typename detail::enable_if_t::value>> + typename std::enable_if_t::value>> : public atomic_ref_base { public: diff --git a/sycl/include/sycl/backend/opencl.hpp b/sycl/include/sycl/backend/opencl.hpp index f60dddd4b1d8d..bbf06d3a1eddc 100644 --- a/sycl/include/sycl/backend/opencl.hpp +++ b/sycl/include/sycl/backend/opencl.hpp @@ -33,7 +33,7 @@ __SYCL_EXPORT bool has_extension(const sycl::device &SyclDevice, const std::string &Extension); // Construction of SYCL platform. -template ::value> * = nullptr> __SYCL_DEPRECATED("Use SYCL 2020 sycl::make_platform free function") T make(typename detail::interop::type Interop) { @@ -41,24 +41,24 @@ T make(typename detail::interop::type Interop) { } // Construction of SYCL device. -template ::value> * = nullptr> +template ::value> * = nullptr> __SYCL_DEPRECATED("Use SYCL 2020 sycl::make_device free function") T make(typename detail::interop::type Interop) { return make_device(detail::pi::cast(Interop)); } // Construction of SYCL context. -template ::value> * = nullptr> +template ::value> + * = nullptr> __SYCL_DEPRECATED("Use SYCL 2020 sycl::make_context free function") T make(typename detail::interop::type Interop) { return make_context(detail::pi::cast(Interop)); } // Construction of SYCL queue. -template ::value> * = nullptr> +template ::value> * = nullptr> __SYCL_DEPRECATED("Use SYCL 2020 sycl::make_queue free function") T make(const context &Context, typename detail::interop::type Interop) { diff --git a/sycl/include/sycl/buffer.hpp b/sycl/include/sycl/buffer.hpp index ffcfcc9802677..2b93734e55532 100644 --- a/sycl/include/sycl/buffer.hpp +++ b/sycl/include/sycl/buffer.hpp @@ -132,8 +132,8 @@ class __SYCL_EXPORT buffer_plain { /// \ingroup sycl_api template >, - typename __Enabled = typename detail::enable_if_t<(dimensions > 0) && - (dimensions <= 3)>> + typename __Enabled = + typename std::enable_if_t<(dimensions > 0) && (dimensions <= 3)>> class buffer : public detail::buffer_plain { // TODO check is_device_copyable::value after converting sycl::vec into a // trivially copyable class. @@ -146,21 +146,21 @@ class buffer : public detail::buffer_plain { using const_reference = const value_type &; using allocator_type = AllocatorT; template - using EnableIfOneDimension = typename detail::enable_if_t<1 == dims>; + using EnableIfOneDimension = typename std::enable_if_t<1 == dims>; // using same requirement for contiguous container as std::span template using EnableIfContiguous = - detail::void_t().data())> (*)[], const T (*)[]>::value>, decltype(std::declval().size())>; template - using EnableIfItInputIterator = detail::enable_if_t< + using EnableIfItInputIterator = std::enable_if_t< std::is_convertible::iterator_category, std::input_iterator_tag>::value>; template - using EnableIfSameNonConstIterators = typename detail::enable_if_t< + using EnableIfSameNonConstIterators = typename std::enable_if_t< std::is_same::value && !std::is_const::value, ItA>; std::array rangeToArray(range<3> &r) { return {r[0], r[1], r[2]}; } @@ -570,7 +570,7 @@ class buffer : public detail::buffer_plain { } template