From 9a1efa714054edec22efc7c8c83f79ba01beb9a2 Mon Sep 17 00:00:00 2001 From: Jakub Stepnowski Date: Wed, 21 Oct 2026 08:42:08 +0200 Subject: [PATCH 01/11] WIP --- sycl/include/sycl/access/access.hpp | 11 ++++ sycl/include/sycl/accessor.hpp | 55 +++++++++++++++++++ sycl/include/sycl/aliases.hpp | 2 + sycl/include/sycl/backend_types.hpp | 2 + .../sycl/detail/address_space_cast.hpp | 2 + sycl/include/sycl/detail/helpers.hpp | 4 ++ sycl/include/sycl/detail/image_ocl_types.hpp | 3 + sycl/include/sycl/group.hpp | 2 + sycl/include/sycl/handler.hpp | 18 ++++++ sycl/include/sycl/image.hpp | 34 ++++++++++++ sycl/include/sycl/multi_ptr.hpp | 34 ++++++++++++ sycl/include/sycl/nd_item.hpp | 8 +++ sycl/include/sycl/sampler.hpp | 2 + sycl/include/sycl/sycl_span.hpp | 8 +++ .../Basic/accessor/Inputs/device_accessor.cpp | 8 +++ 15 files changed, 193 insertions(+) diff --git a/sycl/include/sycl/access/access.hpp b/sycl/include/sycl/access/access.hpp index b2c2eb3957b09..58ecd6a6e46a7 100644 --- a/sycl/include/sycl/access/access.hpp +++ b/sycl/include/sycl/access/access.hpp @@ -106,8 +106,11 @@ template struct mode_target_tag_t { inline constexpr mode_tag_t read_only{}; inline constexpr mode_tag_t read_write{}; inline constexpr mode_tag_t write_only{}; +__SYCL_SUPPRESS_DEPRECATED_PUSH +__SYCL2020_DEPRECATED("use 'read_only' instead") inline constexpr mode_target_tag_t read_constant{}; +__SYCL_SUPPRESS_DEPRECATED_POP inline constexpr mode_target_tag_t read_only_host_task; inline constexpr mode_target_tag_t @@ -118,12 +121,16 @@ inline constexpr mode_target_tag_t namespace detail { constexpr bool isTargetHostAccess(access::target T) { +__SYCL_SUPPRESS_DEPRECATED_PUSH return T == access::target::host_buffer || T == access::target::host_image; +__SYCL_SUPPRESS_DEPRECATED_POP } constexpr bool modeNeedsOldData(access::mode m) { +__SYCL_SUPPRESS_DEPRECATED_PUSH return m == access::mode::read || m == access::mode::write || m == access::mode::read_write || m == access::mode::atomic; +__SYCL_SUPPRESS_DEPRECATED_POP } constexpr bool modeWritesNewData(access::mode m) { @@ -161,8 +168,10 @@ template <> struct TargetToAS { }; template <> struct TargetToAS { +__SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static access::address_space AS = access::address_space::constant_space; +__SYCL_SUPPRESS_DEPRECATED_POP }; template @@ -227,8 +236,10 @@ template struct deduce_AS_impl<__OPENCL_LOCAL_AS__ T> { }; template struct deduce_AS_impl<__OPENCL_CONSTANT_AS__ T> { +__SYCL_SUPPRESS_DEPRECATED_PUSH static constexpr access::address_space value = access::address_space::constant_space; +__SYCL_SUPPRESS_DEPRECATED_POP }; template diff --git a/sycl/include/sycl/accessor.hpp b/sycl/include/sycl/accessor.hpp index 0d8ea29488948..6346c81a4b3c4 100644 --- a/sycl/include/sycl/accessor.hpp +++ b/sycl/include/sycl/accessor.hpp @@ -290,7 +290,9 @@ class accessor_common { protected: constexpr static access::address_space AS = TargetToAS::AS; +__SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsHostBuf = AccessTarget == access::target::host_buffer; +__SYCL_SUPPRESS_DEPRECATED_POP constexpr static bool IsHostTask = AccessTarget == access::target::host_task; // SYCL2020 4.7.6.9.4.3 // IsPlaceHolder template parameter has no bearing on whether the accessor @@ -307,17 +309,23 @@ class accessor_common { // TODO: SYCL 2020 deprecates four of the target enum values // and replaces them with 2 (device and host_task). May want // to change these constexpr. +__SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsGlobalBuf = AccessTarget == access::target::global_buffer; +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsConstantBuf = AccessTarget == access::target::constant_buffer; +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsAccessAnyWrite = AccessMode == access::mode::write || AccessMode == access::mode::read_write || AccessMode == access::mode::discard_write || AccessMode == access::mode::discard_read_write; +__SYCL_SUPPRESS_DEPRECATED_POP constexpr static bool IsAccessReadOnly = AccessMode == access::mode::read; static constexpr bool IsConst = std::is_const_v; @@ -325,7 +333,9 @@ class accessor_common { constexpr static bool IsAccessReadWrite = AccessMode == access::mode::read_write; +__SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsAccessAtomic = AccessMode == access::mode::atomic; +__SYCL_SUPPRESS_DEPRECATED_POP using RefType = detail::const_if_const_AS &; using ConstRefType = const DataT &; @@ -398,6 +408,7 @@ constexpr access::mode deduceAccessMode() { return access::mode::write; } +__SYCL_SUPPRESS_DEPRECATED_PUSH if constexpr (std::is_same_v< MayBeTag1, mode_target_tag_t>) { return access::mode::read; } +__SYCL_SUPPRESS_DEPRECATED_POP if constexpr (std::is_same_v constexpr access::target deduceAccessTarget(access::target defaultTarget) { if constexpr (std::is_same_v< @@ -442,6 +455,7 @@ constexpr access::target deduceAccessTarget(access::target defaultTarget) { access::target::constant_buffer>>) { return access::target::constant_buffer; } +__SYCL_SUPPRESS_DEPRECATED_POP if constexpr ( std::is_same_v::value, "PropertyListT must be accessor_property_list"); @@ -679,6 +697,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : static access::mode getAdjustedMode(const PropertyListT &PropertyList) { access::mode AdjustedMode = AccessMode; +__SYCL_SUPPRESS_DEPRECATED_PUSH if (PropertyList.template has_property() || PropertyList.template has_property()) { if (AdjustedMode == access::mode::write) { @@ -686,6 +705,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : } else if (AdjustedMode == access::mode::read_write) { AdjustedMode = access::mode::discard_read_write; } +__SYCL_SUPPRESS_DEPRECATED_POP } return AdjustedMode; @@ -1666,6 +1686,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : return *(getQualifiedPtr() + LinearIndex); } +__SYCL_SUPPRESS_DEPRECATED_PUSH template > @@ -1673,7 +1694,9 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : *getQualifiedPtr() = Other; return *this; } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template > @@ -1681,6 +1704,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : *getQualifiedPtr() = std::move(Other); return *this; } +__SYCL_SUPPRESS_DEPRECATED_POP template 0) && @@ -1690,6 +1714,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : return getQualifiedPtr()[LinearIndex]; } +__SYCL_SUPPRESS_DEPRECATED_PUSH template operator typename std::enable_if_t< Dims == 0 && AccessMode == access::mode::atomic, atomic>() @@ -1698,7 +1723,9 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t<(Dims > 0) && AccessMode == access::mode::atomic, atomic> @@ -1707,7 +1734,9 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t> @@ -1716,11 +1745,13 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } +__SYCL_SUPPRESS_DEPRECATED_POP template 1)>> auto operator[](size_t Index) const { return AccessorSubscript(*this, Index); } +__SYCL_SUPPRESS_DEPRECATED_PUSH template get_pointer() const noexcept { return getPointerAdjusted(); } +__SYCL_SUPPRESS_DEPRECATED_POP template < access::target AccessTarget_ = AccessTarget, @@ -1740,12 +1772,14 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : getPointerAdjusted())); } +__SYCL_SUPPRESS_DEPRECATED_PUSH template > constant_ptr get_pointer() const { return constant_ptr(getPointerAdjusted()); } +__SYCL_SUPPRESS_DEPRECATED_POP template typename std::enable_if_t< @@ -2055,6 +2090,7 @@ accessor(buffer, handler &, Type1, Type2, Type3, /// Local accessor /// /// \ingroup sycl_api_acc +__SYCL_SUPPRESS_DEPRECATED_PUSH template class __SYCL_SPECIAL_CLASS local_accessor_base : @@ -2063,12 +2099,15 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : #endif public detail::accessor_common { +__SYCL_SUPPRESS_DEPRECATED_POP protected: constexpr static int AdjustedDim = Dimensions == 0 ? 1 : Dimensions; +__SYCL_SUPPRESS_DEPRECATED_PUSH using AccessorCommonT = detail::accessor_common; +__SYCL_SUPPRESS_DEPRECATED_POP using AccessorCommonT::AS; @@ -2274,6 +2313,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : return getQualifiedPtr()[Index]; } +__SYCL_SUPPRESS_DEPRECATED_PUSH template operator typename std::enable_if_t< Dims == 0 && AccessMode == access::mode::atomic, atomic>() @@ -2281,7 +2321,9 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : return atomic( multi_ptr(getQualifiedPtr())); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t<(Dims > 0) && AccessMode == access::mode::atomic, atomic> @@ -2290,7 +2332,9 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t> @@ -2298,6 +2342,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : return atomic(multi_ptr( getQualifiedPtr() + Index)); } +__SYCL_SUPPRESS_DEPRECATED_POP template 1)>> typename AccessorCommonT::template AccessorSubscript< @@ -2315,6 +2360,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : } }; +__SYCL_SUPPRESS_DEPRECATED_PUSH // TODO: Remove deprecated specialization once no longer needed template @@ -2324,6 +2370,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS accessor< public detail::OwnerLessBase< accessor> { +__SYCL_SUPPRESS_DEPRECATED_POP using local_acc = local_accessor_base; @@ -2547,6 +2594,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(local_accessor) local_accessor friend class ext::oneapi::experimental::dynamic_local_accessor; }; +__SYCL_SUPPRESS_DEPRECATED_PUSH template class __SYCL_EBO host_accessor @@ -2557,10 +2605,12 @@ class __SYCL_EBO host_accessor protected: using AccessorT = accessor; +__SYCL_SUPPRESS_DEPRECATED_POP constexpr static int AdjustedDim = Dimensions == 0 ? 1 : Dimensions; constexpr static bool IsAccessReadOnly = AccessMode == access::mode::read; +__SYCL_SUPPRESS_DEPRECATED_PUSH template struct IsSameAsBuffer : std::bool_constant && (Dims > 0) && @@ -2571,6 +2621,7 @@ class __SYCL_EBO host_accessor access::placeholder::false_t>::ConcreteASPtrType Ptr, range AccessRange, range MemRange, id Offset) { +__SYCL_SUPPRESS_DEPRECATED_POP AccessorT::__init(Ptr, AccessRange, MemRange, Offset); } @@ -2721,19 +2772,23 @@ class __SYCL_EBO host_accessor : host_accessor(BufferRef, CommandGroupHandler, AccessRange, AccessOffset, PropertyList, CodeLoc) {} +__SYCL_SUPPRESS_DEPRECATED_PUSH template > const host_accessor & operator=(const typename AccessorT::value_type &Other) const { +__SYCL_SUPPRESS_DEPRECATED_POP *AccessorT::getQualifiedPtr() = Other; return *this; } +__SYCL_SUPPRESS_DEPRECATED_PUSH template > const host_accessor &operator=(typename AccessorT::value_type &&Other) const { +__SYCL_SUPPRESS_DEPRECATED_POP *AccessorT::getQualifiedPtr() = std::move(Other); return *this; } diff --git a/sycl/include/sycl/aliases.hpp b/sycl/include/sycl/aliases.hpp index 828873af46e01..82a2af9162ab7 100644 --- a/sycl/include/sycl/aliases.hpp +++ b/sycl/include/sycl/aliases.hpp @@ -137,11 +137,13 @@ using cl_float = float; using cl_double = double; } // namespace opencl +__SYCL_SUPPRESS_DEPRECATED_PUSH __SYCL_2020_MAKE_VECTOR_ALIASES_FOR_VECTOR_LENGTH(2) __SYCL_2020_MAKE_VECTOR_ALIASES_FOR_VECTOR_LENGTH(3) __SYCL_2020_MAKE_VECTOR_ALIASES_FOR_VECTOR_LENGTH(4) __SYCL_2020_MAKE_VECTOR_ALIASES_FOR_VECTOR_LENGTH(8) __SYCL_2020_MAKE_VECTOR_ALIASES_FOR_VECTOR_LENGTH(16) +__SYCL_SUPPRESS_DEPRECATED_POP } // namespace _V1 } // namespace sycl diff --git a/sycl/include/sycl/backend_types.hpp b/sycl/include/sycl/backend_types.hpp index e80a01bcf1d78..e0349f327a116 100644 --- a/sycl/include/sycl/backend_types.hpp +++ b/sycl/include/sycl/backend_types.hpp @@ -44,8 +44,10 @@ __SYCL_EXPORT std::ostream &operator<<(std::ostream &Out, backend be); namespace detail { inline std::string_view get_backend_name_no_vendor(backend Backend) { switch (Backend) { +__SYCL_SUPPRESS_DEPRECATED_PUSH case backend::host: return "host"; +__SYCL_SUPPRESS_DEPRECATED_POP case backend::opencl: return "opencl"; case backend::ext_oneapi_level_zero: diff --git a/sycl/include/sycl/detail/address_space_cast.hpp b/sycl/include/sycl/detail/address_space_cast.hpp index 95a48e6474cd9..a44ed3256548f 100644 --- a/sycl/include/sycl/detail/address_space_cast.hpp +++ b/sycl/include/sycl/detail/address_space_cast.hpp @@ -22,7 +22,9 @@ inline constexpr bool address_space_cast_is_possible(access::address_space Src, access::address_space Dst) { // constant_space is unique and is not interchangeable with any other. +__SYCL_SUPPRESS_DEPRECATED_PUSH auto constant_space = access::address_space::constant_space; +__SYCL_SUPPRESS_DEPRECATED_POP if (Src == constant_space || Dst == constant_space) return Src == Dst; diff --git a/sycl/include/sycl/detail/helpers.hpp b/sycl/include/sycl/detail/helpers.hpp index 9110b32584a00..2478bd9b49e3a 100644 --- a/sycl/include/sycl/detail/helpers.hpp +++ b/sycl/include/sycl/detail/helpers.hpp @@ -186,10 +186,12 @@ getSPIRVMemorySemanticsMask(memory_order) { } #ifndef __INTEL_PREVIEW_BREAKING_CHANGES +__SYCL_SUPPRESS_DEPRECATED_PUSH inline constexpr uint32_t getSPIRVMemorySemanticsMask(const access::fence_space AccessSpace, const __spv::MemorySemanticsMask LocalScopeMask = __spv::MemorySemanticsMask::WorkgroupMemory) { +__SYCL_SUPPRESS_DEPRECATED_POP // Huge ternary operator below is a workaround for constexpr function // requirement that such function can only contain return statement and // nothing more @@ -215,6 +217,7 @@ getSPIRVMemorySemanticsMask(const access::fence_space AccessSpace, // } // return Flags; +__SYCL_SUPPRESS_DEPRECATED_PUSH return (AccessSpace == access::fence_space::global_space) ? static_cast( __spv::MemorySemanticsMask::SequentiallyConsistent | @@ -229,6 +232,7 @@ getSPIRVMemorySemanticsMask(const access::fence_space AccessSpace, __spv::MemorySemanticsMask::SequentiallyConsistent | __spv::MemorySemanticsMask::CrossWorkgroupMemory | LocalScopeMask); +__SYCL_SUPPRESS_DEPRECATED_POP } #endif // __INTEL_PREVIEW_BREAKING_CHANGES diff --git a/sycl/include/sycl/detail/image_ocl_types.hpp b/sycl/include/sycl/detail/image_ocl_types.hpp index 04329bea7c96b..6524278ecf631 100644 --- a/sycl/include/sycl/detail/image_ocl_types.hpp +++ b/sycl/include/sycl/detail/image_ocl_types.hpp @@ -552,6 +552,7 @@ inline int getSPIRVElementSize(int ImageChannelType, int ImageChannelOrder) { template struct opencl_image_type; +__SYCL_SUPPRESS_DEPRECATED_PUSH // Creation of dummy ocl types for host_image targets. // These dummy ocl types are needed by the compiler parser for the compilation // of host application code with __SYCL_DEVICE_ONLY__ macro set. @@ -560,6 +561,8 @@ struct opencl_image_type { using type = opencl_image_type *; }; +__SYCL_SUPPRESS_DEPRECATED_PUSH + template struct sampled_opencl_image_type { using type = void *; }; diff --git a/sycl/include/sycl/group.hpp b/sycl/include/sycl/group.hpp index c5f93cd5fe957..d09eedd7d4515 100644 --- a/sycl/include/sycl/group.hpp +++ b/sycl/include/sycl/group.hpp @@ -307,6 +307,7 @@ template class __SYCL_TYPE(group) group { } #ifndef __INTEL_PREVIEW_BREAKING_CHANGES +__SYCL_SUPPRESS_DEPRECATED_PUSH /// Executes a work-group mem-fence with memory ordering on the local address /// space, global address space or both based on the value of \p accessSpace. template @@ -318,6 +319,7 @@ template class __SYCL_TYPE(group) group { accessMode == access::mode::read_write, access::fence_space> accessSpace = access::fence_space::global_and_local) const { +__SYCL_SUPPRESS_DEPRECATED_POP #ifdef __SYCL_DEVICE_ONLY__ uint32_t flags = detail::getSPIRVMemorySemanticsMask(accessSpace); // TODO: currently, there is no good way in SPIR-V to set the memory diff --git a/sycl/include/sycl/handler.hpp b/sycl/include/sycl/handler.hpp index 56539906b52d3..94096291bbe16 100644 --- a/sycl/include/sycl/handler.hpp +++ b/sycl/include/sycl/handler.hpp @@ -473,6 +473,7 @@ class __SYCL_EXPORT handler { void setArgsHelper(int) {} +__SYCL_SUPPRESS_DEPRECATED_PUSH void setLocalAccessorArgHelper(int ArgIndex, detail::LocalAccessorBaseHost &LocalAccBase) { detail::LocalAccessorImplPtr LocalAccImpl = @@ -482,13 +483,16 @@ class __SYCL_EXPORT handler { addArg(detail::kernel_param_kind_t::kind_accessor, Req, static_cast(access::target::local), ArgIndex); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH // setArgHelper for local accessor argument (legacy accessor interface) template void setArgHelper(int ArgIndex, accessor &&Arg) { +__SYCL_SUPPRESS_DEPRECATED_POP (void)ArgIndex; (void)Arg; #ifndef __SYCL_DEVICE_ONLY__ @@ -508,6 +512,7 @@ class __SYCL_EXPORT handler { void setArgHelper(int ArgIndex, detail::work_group_memory_impl &Arg); +__SYCL_SUPPRESS_DEPRECATED_PUSH // setArgHelper for non local accessor argument. template @@ -515,6 +520,7 @@ class __SYCL_EXPORT handler { setArgHelper( int ArgIndex, accessor &&Arg) { +__SYCL_SUPPRESS_DEPRECATED_POP detail::AccessorBaseHost *AccBase = (detail::AccessorBaseHost *)&Arg; const detail::AccessorImplPtr &AccImpl = detail::getSyclObjImpl(*AccBase); // Ensure the data of AccImpl lives at least as long as the handler. @@ -541,11 +547,13 @@ class __SYCL_EXPORT handler { } } +__SYCL_SUPPRESS_DEPRECATED_PUSH void setArgHelper(int ArgIndex, sampler &&Arg) { void *StoredArg = storePlainArg(Arg); addArg(detail::kernel_param_kind_t::kind_sampler, StoredArg, sizeof(sampler), ArgIndex); } +__SYCL_SUPPRESS_DEPRECATED_POP void setArgHelper(int ArgIndex, stream &&Str); @@ -763,15 +771,19 @@ class __SYCL_EXPORT handler { return false; } +__SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool isConstOrGlobal(access::target AccessTarget) { return AccessTarget == access::target::device || AccessTarget == access::target::constant_buffer; } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool isImageOrImageArray(access::target AccessTarget) { return AccessTarget == access::target::image || AccessTarget == access::target::image_array; } +__SYCL_SUPPRESS_DEPRECATED_POP constexpr static bool isValidTargetForExplicitOp(access::target AccessTarget) { @@ -783,6 +795,7 @@ class __SYCL_EXPORT handler { AccessMode == access::mode::read_write; } +__SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool isValidModeForDestinationAccessor(access::mode AccessMode) { return AccessMode == access::mode::write || @@ -790,6 +803,7 @@ class __SYCL_EXPORT handler { AccessMode == access::mode::discard_write || AccessMode == access::mode::discard_read_write; } +__SYCL_SUPPRESS_DEPRECATED_POP // UR APIs only support select fill sizes: 1, 2, 4, 8, 16, 32, 64, 128 constexpr static bool isBackendSupportedFillSize(size_t Size) { @@ -1165,6 +1179,7 @@ class __SYCL_EXPORT handler { template using is_same_type = std::is_same, remove_cv_ref_t>; +__SYCL_SUPPRESS_DEPRECATED_PUSH template struct ShouldEnableSetArg { static constexpr bool value = std::is_trivially_copyable_v> @@ -1178,6 +1193,7 @@ class __SYCL_EXPORT handler { || is_same_type::value // Stream || detail::check_if_device_copyable_v>; }; +__SYCL_SUPPRESS_DEPRECATED_POP /// Sets argument for OpenCL interoperability kernels. /// @@ -1213,7 +1229,9 @@ class __SYCL_EXPORT handler { constexpr int AccessTargetMask = 0x7ff; const access::target target = static_cast( type::sizes[NumArgs] & AccessTargetMask); +__SYCL_SUPPRESS_DEPRECATED_PUSH if (target == target::local) { +__SYCL_SUPPRESS_DEPRECATED_POP detail::LocalAccessorBaseHost *LocalAccBase = (detail::LocalAccessorBaseHost *)(FieldArg); setLocalAccessorArgHelper(ArgIndex + NumArgs, *LocalAccBase); diff --git a/sycl/include/sycl/image.hpp b/sycl/include/sycl/image.hpp index a43db0f08c4be..d3c4aab2706d4 100644 --- a/sycl/include/sycl/image.hpp +++ b/sycl/include/sycl/image.hpp @@ -107,7 +107,9 @@ enum class image_format : unsigned int { using byte = unsigned char; +__SYCL_SUPPRESS_DEPRECATED_PUSH using image_allocator = detail::aligned_allocator; +__SYCL_SUPPRESS_DEPRECATED_POP namespace detail { @@ -759,6 +761,7 @@ class unsampled_image #endif public: +__SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( image_format Format, const range &Range, const property_list &PropList = {}, @@ -773,7 +776,9 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( image_format Format, const range &Range, AllocatorT Allocator, const property_list &PropList = {}, @@ -789,7 +794,9 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -807,7 +814,9 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -827,7 +836,9 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( void *HostPointer, image_format Format, const range &Range, const property_list &PropList = {}, @@ -842,7 +853,9 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( void *HostPointer, image_format Format, const range &Range, AllocatorT Allocator, const property_list &PropList = {}, @@ -858,7 +871,9 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -876,7 +891,9 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -896,7 +913,9 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( std::shared_ptr &HostPointer, image_format Format, const range &Range, const property_list &PropList = {}, @@ -911,7 +930,9 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( std::shared_ptr &HostPointer, image_format Format, const range &Range, AllocatorT Allocator, @@ -928,7 +949,9 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -947,7 +970,9 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -967,6 +992,7 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format); } +__SYCL_SUPPRESS_DEPRECATED_POP #ifdef __DPCPP_ENABLE_UNFINISHED_KHR_EXTENSIONS template &Range, const property_list &PropList = {}, @@ -1187,7 +1214,9 @@ class sampled_image CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format, Sampler); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> sampled_image( @@ -1206,7 +1235,9 @@ class sampled_image CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format, Sampler); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH sampled_image( std::shared_ptr &HostPointer, image_format Format, image_sampler Sampler, const range &Range, @@ -1222,7 +1253,9 @@ class sampled_image CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format, Sampler); } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> sampled_image( @@ -1241,6 +1274,7 @@ class sampled_image CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format, Sampler); } +__SYCL_SUPPRESS_DEPRECATED_POP #ifdef __DPCPP_ENABLE_UNFINISHED_KHR_EXTENSIONS template struct LegacyPointerTypes { +__SYCL_SUPPRESS_DEPRECATED_PUSH using decorated_type = typename detail::DecoratedType< ElementType, access::address_space::constant_space>::type; +__SYCL_SUPPRESS_DEPRECATED_POP using pointer_t = decorated_type *; using const_pointer_t = decorated_type const *; }; @@ -60,8 +62,10 @@ struct LegacyReferenceTypes { template struct LegacyReferenceTypes { +__SYCL_SUPPRESS_DEPRECATED_PUSH using decorated_type = typename detail::DecoratedType< ElementType, access::address_space::constant_space>::type; +__SYCL_SUPPRESS_DEPRECATED_POP using reference_t = decorated_type &; using const_reference_t = decorated_type &; }; @@ -258,6 +262,7 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { m_Pointer = nullptr; return *this; } +__SYCL_SUPPRESS_DEPRECATED_PUSH template < access::address_space OtherSpace, access::decorated OtherIsDecorated, typename = @@ -268,6 +273,8 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { m_Pointer = detail::static_address_cast(Other.get_decorated()); return *this; } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template < access::address_space OtherSpace, access::decorated OtherIsDecorated, typename = @@ -278,6 +285,7 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { m_Pointer = detail::static_address_cast(std::move(Other.m_Pointer)); return *this; } +__SYCL_SUPPRESS_DEPRECATED_POP reference operator*() const { return *m_Pointer; } pointer operator->() const { return get(); } @@ -563,9 +571,11 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { // Legacy has a different interface. static_assert(DecorateAddress != access::decorated::legacy); // constant_space is only supported in legacy multi_ptr. +__SYCL_SUPPRESS_DEPRECATED_PUSH static_assert(Space != access::address_space::constant_space, "SYCL 2020 multi_ptr does not support the deprecated " "constant_space address space."); +__SYCL_SUPPRESS_DEPRECATED_POP // Constructors multi_ptr() : m_Pointer(nullptr) {} @@ -813,6 +823,7 @@ class multi_ptr { Accessor.get_pointer().get())) {} // Only if Space == local_space || generic_space +__SYCL_SUPPRESS_DEPRECATED_PUSH template < int dimensions, access::mode Mode, access::placeholder isPlaceholder, typename PropertyListT, access::address_space _Space = Space, @@ -823,6 +834,7 @@ class multi_ptr { isPlaceholder, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} +__SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == local_space || generic_space template @@ -830,6 +842,7 @@ class multi_ptr { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == constant_space +__SYCL_SUPPRESS_DEPRECATED_PUSH template < int dimensions, access::mode Mode, access::placeholder isPlaceholder, typename PropertyListT, access::address_space _Space = Space, @@ -839,6 +852,7 @@ class multi_ptr { isPlaceholder, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} +__SYCL_SUPPRESS_DEPRECATED_POP // The following constructors are necessary to create multi_ptr from @@ -867,6 +881,7 @@ class multi_ptr { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == local_space || generic_space and element type is const +__SYCL_SUPPRESS_DEPRECATED_PUSH template { target::local, isPlaceholder, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} +__SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == local_space || generic_space and element type is const template { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == constant_space and element type is const +__SYCL_SUPPRESS_DEPRECATED_PUSH template < int dimensions, access::mode Mode, access::placeholder isPlaceholder, typename PropertyListT, access::address_space _Space = Space, @@ -904,6 +921,7 @@ class multi_ptr { target::constant_buffer, isPlaceholder, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} +__SYCL_SUPPRESS_DEPRECATED_POP // TODO: This constructor is the temporary solution for the existing problem // with conversions from multi_ptr { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == local_space || generic_space +__SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::mode Mode, typename PropertyListT, access::address_space _Space = Space, @@ -1120,8 +1139,10 @@ class multi_ptr { access::placeholder::false_t, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} +__SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == local_space || generic_space +__SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::address_space _Space = Space, @@ -1130,8 +1151,10 @@ class multi_ptr { Space == access::address_space::local_space)>> multi_ptr(local_accessor Accessor) : multi_ptr(Accessor.get_pointer()) {} +__SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == constant_space +__SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::mode Mode, typename PropertyListT, access::address_space _Space = Space, @@ -1141,6 +1164,7 @@ class multi_ptr { access::placeholder::false_t, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} +__SYCL_SUPPRESS_DEPRECATED_POP using ReturnPtr = detail::const_if_const_AS *; // Returns the underlying OpenCL C pointer @@ -1269,6 +1293,7 @@ class multi_ptr { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == local_space || generic_space +__SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::mode Mode, typename PropertyListT, access::address_space _Space = Space, @@ -1279,8 +1304,10 @@ class multi_ptr { access::placeholder::false_t, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} +__SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == local_space || generic_space +__SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::address_space _Space = Space, @@ -1289,8 +1316,10 @@ class multi_ptr { Space == access::address_space::local_space)>> multi_ptr(local_accessor Accessor) : multi_ptr(Accessor.get_pointer()) {} +__SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == constant_space +__SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::mode Mode, typename PropertyListT, access::address_space _Space = Space, @@ -1300,6 +1329,7 @@ class multi_ptr { access::placeholder::false_t, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} +__SYCL_SUPPRESS_DEPRECATED_POP // Returns the underlying OpenCL C pointer pointer_t get() const { return m_Pointer; } @@ -1348,17 +1378,21 @@ template ) -> multi_ptr; +__SYCL_SUPPRESS_DEPRECATED_PUSH template multi_ptr(accessor) -> multi_ptr; +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH template multi_ptr( accessor) -> multi_ptr; +__SYCL_SUPPRESS_DEPRECATED_POP template multi_ptr(local_accessor) -> multi_ptr; diff --git a/sycl/include/sycl/nd_item.hpp b/sycl/include/sycl/nd_item.hpp index 2adc9ecd6b122..052492ab9c647 100644 --- a/sycl/include/sycl/nd_item.hpp +++ b/sycl/include/sycl/nd_item.hpp @@ -71,7 +71,9 @@ template class nd_item { size_t LinId = 0; id Index = get_global_id(); range Extent = get_global_range(); +__SYCL_SUPPRESS_DEPRECATED_PUSH id Offset = get_offset(); +__SYCL_SUPPRESS_DEPRECATED_POP if (1 == Dimensions) { LinId = Index[0] - Offset[0]; } else if (2 == Dimensions) { @@ -198,14 +200,18 @@ template class nd_item { } nd_range get_nd_range() const noexcept { +__SYCL_SUPPRESS_DEPRECATED_PUSH return nd_range(get_global_range(), get_local_range(), get_offset()); +__SYCL_SUPPRESS_DEPRECATED_POP } #ifndef __INTEL_PREVIEW_BREAKING_CHANGES +__SYCL_SUPPRESS_DEPRECATED_PUSH __SYCL2020_DEPRECATED("use sycl::group_barrier() free function instead") void barrier([[maybe_unused]] access::fence_space accessSpace = access::fence_space::global_and_local) const { +__SYCL_SUPPRESS_DEPRECATED_POP #ifdef __SYCL_DEVICE_ONLY__ uint32_t flags = _V1::detail::getSPIRVMemorySemanticsMask(accessSpace); __spirv_ControlBarrier(__spv::Scope::Workgroup, __spv::Scope::Workgroup, @@ -215,6 +221,7 @@ template class nd_item { /// Executes a work-group mem-fence with memory ordering on the local address /// space, global address space or both based on the value of \p accessSpace. +__SYCL_SUPPRESS_DEPRECATED_PUSH template __SYCL2020_DEPRECATED("use sycl::atomic_fence() free function instead") void mem_fence( @@ -224,6 +231,7 @@ template class nd_item { accessMode == access::mode::read_write, access::fence_space> accessSpace = access::fence_space::global_and_local) const { +__SYCL_SUPPRESS_DEPRECATED_POP #if __SYCL_DEVICE_ONLY__ uint32_t flags = detail::getSPIRVMemorySemanticsMask(accessSpace); // TODO: currently, there is no good way in SPIR-V to set the memory diff --git a/sycl/include/sycl/sampler.hpp b/sycl/include/sycl/sampler.hpp index 9d77a6eb7501b..813473a478b6b 100644 --- a/sycl/include/sycl/sampler.hpp +++ b/sycl/include/sycl/sampler.hpp @@ -141,7 +141,9 @@ struct image_sampler { } // namespace _V1 } // namespace sycl +__SYCL_SUPPRESS_DEPRECATED_PUSH template <> struct std::hash : public sycl::detail::sycl_obj_hash {}; +__SYCL_SUPPRESS_DEPRECATED_POP diff --git a/sycl/include/sycl/sycl_span.hpp b/sycl/include/sycl/sycl_span.hpp index 55126d0323e4f..23ca7b4722ce8 100644 --- a/sycl/include/sycl/sycl_span.hpp +++ b/sycl/include/sycl/sycl_span.hpp @@ -393,17 +393,21 @@ template class _SYCL_SPAN_TEMPLATE_VIS span { return rev_iterator(begin()); } +__SYCL_SUPPRESS_DEPRECATED_PUSH _SYCL_SPAN_INLINE_VISIBILITY span __as_bytes() const noexcept { return span{ reinterpret_cast(data()), size_bytes()}; } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH _SYCL_SPAN_INLINE_VISIBILITY span __as_writable_bytes() const noexcept { return span{ reinterpret_cast(data()), size_bytes()}; } +__SYCL_SUPPRESS_DEPRECATED_POP private: pointer __data; @@ -586,15 +590,19 @@ class _SYCL_SPAN_TEMPLATE_VIS span<_Tp, dynamic_extent> { return rev_iterator(begin()); } +__SYCL_SUPPRESS_DEPRECATED_PUSH _SYCL_SPAN_INLINE_VISIBILITY span __as_bytes() const noexcept { return {reinterpret_cast(data()), size_bytes()}; } +__SYCL_SUPPRESS_DEPRECATED_POP +__SYCL_SUPPRESS_DEPRECATED_PUSH _SYCL_SPAN_INLINE_VISIBILITY span __as_writable_bytes() const noexcept { return {reinterpret_cast(data()), size_bytes()}; } +__SYCL_SUPPRESS_DEPRECATED_POP private: pointer __data; diff --git a/sycl/test-e2e/Basic/accessor/Inputs/device_accessor.cpp b/sycl/test-e2e/Basic/accessor/Inputs/device_accessor.cpp index ca51c57755cbd..f50ad95939f8e 100644 --- a/sycl/test-e2e/Basic/accessor/Inputs/device_accessor.cpp +++ b/sycl/test-e2e/Basic/accessor/Inputs/device_accessor.cpp @@ -171,11 +171,13 @@ int main() { sycl::accessor acc_2(buf_data, cgh, sycl::range<1>(8), sycl::no_init); sycl::accessor acc_3(buf_data, cgh, sycl::range<1>(8), sycl::id<1>(1), sycl::no_init); + __SYCL_SUPPRESS_DEPRECATED_PUSH sycl::accessor acc_4(buf_data, cgh, sycl::read_constant); sycl::accessor acc_5(buf_data, cgh, sycl::range<1>(8), sycl::read_constant); sycl::accessor acc_6(buf_data, cgh, sycl::range<1>(8), sycl::id<1>(1), sycl::read_constant); + __SYCL_SUPPRESS_DEPRECATED_POP sycl::accessor acc_7(buf_data, cgh, sycl::write_only, sycl::no_init); sycl::accessor acc_8(buf_data, cgh, sycl::range<1>(8), sycl::write_only, sycl::no_init); @@ -186,11 +188,13 @@ int main() { auto acc_2 = buf_data.get_access(cgh, sycl::range<1>(8), sycl::no_init); auto acc_3 = buf_data.get_access(cgh, sycl::range<1>(8), sycl::id<1>(1), sycl::no_init); + __SYCL_SUPPRESS_DEPRECATED_PUSH auto acc_4 = buf_data.get_access(cgh, sycl::read_constant); auto acc_5 = buf_data.get_access(cgh, sycl::range<1>(8), sycl::read_constant); auto acc_6 = buf_data.get_access(cgh, sycl::range<1>(8), sycl::id<1>(1), sycl::read_constant); + __SYCL_SUPPRESS_DEPRECATED_POP auto acc_7 = buf_data.get_access(cgh, sycl::write_only, sycl::no_init); auto acc_8 = buf_data.get_access(cgh, sycl::range<1>(8), sycl::write_only, sycl::no_init); @@ -241,10 +245,12 @@ int main() { sycl::accessor acc_2(buf_data, sycl::range<1>(8), sycl::no_init); sycl::accessor acc_3(buf_data, sycl::range<1>(8), sycl::id<1>(1), sycl::no_init); + __SYCL_SUPPRESS_DEPRECATED_PUSH sycl::accessor acc_4(buf_data, sycl::read_constant); sycl::accessor acc_5(buf_data, sycl::range<1>(8), sycl::read_constant); sycl::accessor acc_6(buf_data, sycl::range<1>(8), sycl::id<1>(1), sycl::read_constant); + __SYCL_SUPPRESS_DEPRECATED_POP sycl::accessor acc_7(buf_data, sycl::write_only, sycl::no_init); sycl::accessor acc_8(buf_data, sycl::range<1>(8), sycl::write_only, sycl::no_init); @@ -255,10 +261,12 @@ int main() { auto acc_2 = buf_data.get_access(sycl::range<1>(8), sycl::no_init); auto acc_3 = buf_data.get_access(sycl::range<1>(8), sycl::id<1>(1), sycl::no_init); + __SYCL_SUPPRESS_DEPRECATED_PUSH auto acc_4 = buf_data.get_access(sycl::read_constant); auto acc_5 = buf_data.get_access(sycl::range<1>(8), sycl::read_constant); auto acc_6 = buf_data.get_access(sycl::range<1>(8), sycl::id<1>(1), sycl::read_constant); + __SYCL_SUPPRESS_DEPRECATED_POP auto acc_7 = buf_data.get_access(sycl::write_only, sycl::no_init); auto acc_8 = buf_data.get_access(sycl::range<1>(8), sycl::write_only, sycl::no_init); From b9ce104d3113c2f93efe8d684f4c9df8e787c49e Mon Sep 17 00:00:00 2001 From: Jakub Stepnowski Date: Wed, 21 Oct 2026 11:07:42 +0200 Subject: [PATCH 02/11] WIP --- sycl/include/sycl/access/access.hpp | 16 +-- sycl/include/sycl/accessor.hpp | 100 +++++++++--------- sycl/include/sycl/backend_types.hpp | 4 +- .../sycl/detail/address_space_cast.hpp | 4 +- sycl/include/sycl/detail/helpers.hpp | 6 +- sycl/include/sycl/group.hpp | 4 +- sycl/include/sycl/handler.hpp | 37 ++++--- sycl/include/sycl/image.hpp | 64 +++++------ sycl/include/sycl/multi_ptr.hpp | 60 +++++------ sycl/include/sycl/nd_item.hpp | 16 +-- sycl/include/sycl/sycl_span.hpp | 16 +-- 11 files changed, 163 insertions(+), 164 deletions(-) diff --git a/sycl/include/sycl/access/access.hpp b/sycl/include/sycl/access/access.hpp index 58ecd6a6e46a7..a1436e8223527 100644 --- a/sycl/include/sycl/access/access.hpp +++ b/sycl/include/sycl/access/access.hpp @@ -121,16 +121,16 @@ inline constexpr mode_target_tag_t namespace detail { constexpr bool isTargetHostAccess(access::target T) { -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH return T == access::target::host_buffer || T == access::target::host_image; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP } constexpr bool modeNeedsOldData(access::mode m) { -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH return m == access::mode::read || m == access::mode::write || m == access::mode::read_write || m == access::mode::atomic; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP } constexpr bool modeWritesNewData(access::mode m) { @@ -168,10 +168,10 @@ template <> struct TargetToAS { }; template <> struct TargetToAS { -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static access::address_space AS = access::address_space::constant_space; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP }; template @@ -236,10 +236,10 @@ template struct deduce_AS_impl<__OPENCL_LOCAL_AS__ T> { }; template struct deduce_AS_impl<__OPENCL_CONSTANT_AS__ T> { -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH static constexpr access::address_space value = access::address_space::constant_space; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP }; template diff --git a/sycl/include/sycl/accessor.hpp b/sycl/include/sycl/accessor.hpp index 6346c81a4b3c4..e88e66280a5f2 100644 --- a/sycl/include/sycl/accessor.hpp +++ b/sycl/include/sycl/accessor.hpp @@ -290,9 +290,9 @@ class accessor_common { protected: constexpr static access::address_space AS = TargetToAS::AS; -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsHostBuf = AccessTarget == access::target::host_buffer; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP constexpr static bool IsHostTask = AccessTarget == access::target::host_task; // SYCL2020 4.7.6.9.4.3 // IsPlaceHolder template parameter has no bearing on whether the accessor @@ -309,23 +309,23 @@ __SYCL_SUPPRESS_DEPRECATED_POP // TODO: SYCL 2020 deprecates four of the target enum values // and replaces them with 2 (device and host_task). May want // to change these constexpr. -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsGlobalBuf = AccessTarget == access::target::global_buffer; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsConstantBuf = AccessTarget == access::target::constant_buffer; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsAccessAnyWrite = AccessMode == access::mode::write || AccessMode == access::mode::read_write || AccessMode == access::mode::discard_write || AccessMode == access::mode::discard_read_write; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP constexpr static bool IsAccessReadOnly = AccessMode == access::mode::read; static constexpr bool IsConst = std::is_const_v; @@ -333,9 +333,9 @@ __SYCL_SUPPRESS_DEPRECATED_POP constexpr static bool IsAccessReadWrite = AccessMode == access::mode::read_write; -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsAccessAtomic = AccessMode == access::mode::atomic; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP using RefType = detail::const_if_const_AS &; using ConstRefType = const DataT &; @@ -408,7 +408,7 @@ constexpr access::mode deduceAccessMode() { return access::mode::write; } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH if constexpr (std::is_same_v< MayBeTag1, mode_target_tag_t>) { return access::mode::read; } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP if constexpr (std::is_same_v>) { return access::target::constant_buffer; } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP if constexpr ( std::is_same_v::value, "PropertyListT must be accessor_property_list"); @@ -697,7 +697,7 @@ __SYCL_SUPPRESS_DEPRECATED_POP static access::mode getAdjustedMode(const PropertyListT &PropertyList) { access::mode AdjustedMode = AccessMode; -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH if (PropertyList.template has_property() || PropertyList.template has_property()) { if (AdjustedMode == access::mode::write) { @@ -705,8 +705,8 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH } else if (AdjustedMode == access::mode::read_write) { AdjustedMode = access::mode::discard_read_write; } -__SYCL_SUPPRESS_DEPRECATED_POP } + __SYCL_SUPPRESS_DEPRECATED_POP return AdjustedMode; } @@ -1686,7 +1686,7 @@ __SYCL_SUPPRESS_DEPRECATED_POP return *(getQualifiedPtr() + LinearIndex); } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template > @@ -1694,9 +1694,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH *getQualifiedPtr() = Other; return *this; } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template > @@ -1704,7 +1704,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH *getQualifiedPtr() = std::move(Other); return *this; } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP template 0) && @@ -1714,7 +1714,7 @@ __SYCL_SUPPRESS_DEPRECATED_POP return getQualifiedPtr()[LinearIndex]; } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template operator typename std::enable_if_t< Dims == 0 && AccessMode == access::mode::atomic, atomic>() @@ -1723,9 +1723,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t<(Dims > 0) && AccessMode == access::mode::atomic, atomic> @@ -1734,9 +1734,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t> @@ -1745,13 +1745,13 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP template 1)>> auto operator[](size_t Index) const { return AccessorSubscript(*this, Index); } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template get_pointer() const noexcept { return getPointerAdjusted(); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP template < access::target AccessTarget_ = AccessTarget, @@ -1772,14 +1772,14 @@ __SYCL_SUPPRESS_DEPRECATED_POP getPointerAdjusted())); } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template > constant_ptr get_pointer() const { return constant_ptr(getPointerAdjusted()); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP template { -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP protected: constexpr static int AdjustedDim = Dimensions == 0 ? 1 : Dimensions; -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH using AccessorCommonT = detail::accessor_common; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP using AccessorCommonT::AS; @@ -2313,7 +2313,7 @@ __SYCL_SUPPRESS_DEPRECATED_POP return getQualifiedPtr()[Index]; } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template operator typename std::enable_if_t< Dims == 0 && AccessMode == access::mode::atomic, atomic>() @@ -2321,9 +2321,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH return atomic( multi_ptr(getQualifiedPtr())); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t<(Dims > 0) && AccessMode == access::mode::atomic, atomic> @@ -2332,9 +2332,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t> @@ -2342,7 +2342,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH return atomic(multi_ptr( getQualifiedPtr() + Index)); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP template 1)>> typename AccessorCommonT::template AccessorSubscript< @@ -2370,7 +2370,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS accessor< public detail::OwnerLessBase< accessor> { -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP using local_acc = local_accessor_base; @@ -2605,12 +2605,12 @@ class __SYCL_EBO host_accessor protected: using AccessorT = accessor; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP constexpr static int AdjustedDim = Dimensions == 0 ? 1 : Dimensions; constexpr static bool IsAccessReadOnly = AccessMode == access::mode::read; -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template struct IsSameAsBuffer : std::bool_constant && (Dims > 0) && @@ -2621,7 +2621,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH access::placeholder::false_t>::ConcreteASPtrType Ptr, range AccessRange, range MemRange, id Offset) { -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP AccessorT::__init(Ptr, AccessRange, MemRange, Offset); } @@ -2772,23 +2772,23 @@ __SYCL_SUPPRESS_DEPRECATED_POP : host_accessor(BufferRef, CommandGroupHandler, AccessRange, AccessOffset, PropertyList, CodeLoc) {} -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template > const host_accessor & operator=(const typename AccessorT::value_type &Other) const { -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP *AccessorT::getQualifiedPtr() = Other; return *this; } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template > const host_accessor &operator=(typename AccessorT::value_type &&Other) const { -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP *AccessorT::getQualifiedPtr() = std::move(Other); return *this; } diff --git a/sycl/include/sycl/backend_types.hpp b/sycl/include/sycl/backend_types.hpp index e0349f327a116..50fea382a8d5d 100644 --- a/sycl/include/sycl/backend_types.hpp +++ b/sycl/include/sycl/backend_types.hpp @@ -44,10 +44,10 @@ __SYCL_EXPORT std::ostream &operator<<(std::ostream &Out, backend be); namespace detail { inline std::string_view get_backend_name_no_vendor(backend Backend) { switch (Backend) { -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH case backend::host: return "host"; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP case backend::opencl: return "opencl"; case backend::ext_oneapi_level_zero: diff --git a/sycl/include/sycl/detail/address_space_cast.hpp b/sycl/include/sycl/detail/address_space_cast.hpp index a44ed3256548f..4a5a065fe70a0 100644 --- a/sycl/include/sycl/detail/address_space_cast.hpp +++ b/sycl/include/sycl/detail/address_space_cast.hpp @@ -22,9 +22,9 @@ inline constexpr bool address_space_cast_is_possible(access::address_space Src, access::address_space Dst) { // constant_space is unique and is not interchangeable with any other. -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH auto constant_space = access::address_space::constant_space; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP if (Src == constant_space || Dst == constant_space) return Src == Dst; diff --git a/sycl/include/sycl/detail/helpers.hpp b/sycl/include/sycl/detail/helpers.hpp index 2478bd9b49e3a..26bf504041776 100644 --- a/sycl/include/sycl/detail/helpers.hpp +++ b/sycl/include/sycl/detail/helpers.hpp @@ -191,7 +191,7 @@ inline constexpr uint32_t getSPIRVMemorySemanticsMask(const access::fence_space AccessSpace, const __spv::MemorySemanticsMask LocalScopeMask = __spv::MemorySemanticsMask::WorkgroupMemory) { -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP // Huge ternary operator below is a workaround for constexpr function // requirement that such function can only contain return statement and // nothing more @@ -217,7 +217,7 @@ __SYCL_SUPPRESS_DEPRECATED_POP // } // return Flags; -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH return (AccessSpace == access::fence_space::global_space) ? static_cast( __spv::MemorySemanticsMask::SequentiallyConsistent | @@ -232,7 +232,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH __spv::MemorySemanticsMask::SequentiallyConsistent | __spv::MemorySemanticsMask::CrossWorkgroupMemory | LocalScopeMask); -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP } #endif // __INTEL_PREVIEW_BREAKING_CHANGES diff --git a/sycl/include/sycl/group.hpp b/sycl/include/sycl/group.hpp index d09eedd7d4515..e138b27455fb3 100644 --- a/sycl/include/sycl/group.hpp +++ b/sycl/include/sycl/group.hpp @@ -307,7 +307,7 @@ template class __SYCL_TYPE(group) group { } #ifndef __INTEL_PREVIEW_BREAKING_CHANGES -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH /// Executes a work-group mem-fence with memory ordering on the local address /// space, global address space or both based on the value of \p accessSpace. template @@ -319,7 +319,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH accessMode == access::mode::read_write, access::fence_space> accessSpace = access::fence_space::global_and_local) const { -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP #ifdef __SYCL_DEVICE_ONLY__ uint32_t flags = detail::getSPIRVMemorySemanticsMask(accessSpace); // TODO: currently, there is no good way in SPIR-V to set the memory diff --git a/sycl/include/sycl/handler.hpp b/sycl/include/sycl/handler.hpp index 94096291bbe16..31e5b03c462a3 100644 --- a/sycl/include/sycl/handler.hpp +++ b/sycl/include/sycl/handler.hpp @@ -473,7 +473,7 @@ class __SYCL_EXPORT handler { void setArgsHelper(int) {} -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH void setLocalAccessorArgHelper(int ArgIndex, detail::LocalAccessorBaseHost &LocalAccBase) { detail::LocalAccessorImplPtr LocalAccImpl = @@ -483,17 +483,16 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH addArg(detail::kernel_param_kind_t::kind_accessor, Req, static_cast(access::target::local), ArgIndex); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH // setArgHelper for local accessor argument (legacy accessor interface) template void setArgHelper(int ArgIndex, accessor &&Arg) { -__SYCL_SUPPRESS_DEPRECATED_POP - (void)ArgIndex; + __SYCL_SUPPRESS_DEPRECATED_POP(void) ArgIndex; (void)Arg; #ifndef __SYCL_DEVICE_ONLY__ setLocalAccessorArgHelper(ArgIndex, Arg); @@ -512,7 +511,7 @@ __SYCL_SUPPRESS_DEPRECATED_POP void setArgHelper(int ArgIndex, detail::work_group_memory_impl &Arg); -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH // setArgHelper for non local accessor argument. template @@ -520,7 +519,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH setArgHelper( int ArgIndex, accessor &&Arg) { -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP detail::AccessorBaseHost *AccBase = (detail::AccessorBaseHost *)&Arg; const detail::AccessorImplPtr &AccImpl = detail::getSyclObjImpl(*AccBase); // Ensure the data of AccImpl lives at least as long as the handler. @@ -547,13 +546,13 @@ __SYCL_SUPPRESS_DEPRECATED_POP } } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH void setArgHelper(int ArgIndex, sampler &&Arg) { void *StoredArg = storePlainArg(Arg); addArg(detail::kernel_param_kind_t::kind_sampler, StoredArg, sizeof(sampler), ArgIndex); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP void setArgHelper(int ArgIndex, stream &&Str); @@ -771,19 +770,19 @@ __SYCL_SUPPRESS_DEPRECATED_POP return false; } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool isConstOrGlobal(access::target AccessTarget) { return AccessTarget == access::target::device || AccessTarget == access::target::constant_buffer; } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool isImageOrImageArray(access::target AccessTarget) { return AccessTarget == access::target::image || AccessTarget == access::target::image_array; } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP constexpr static bool isValidTargetForExplicitOp(access::target AccessTarget) { @@ -795,7 +794,7 @@ __SYCL_SUPPRESS_DEPRECATED_POP AccessMode == access::mode::read_write; } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool isValidModeForDestinationAccessor(access::mode AccessMode) { return AccessMode == access::mode::write || @@ -803,7 +802,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH AccessMode == access::mode::discard_write || AccessMode == access::mode::discard_read_write; } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP // UR APIs only support select fill sizes: 1, 2, 4, 8, 16, 32, 64, 128 constexpr static bool isBackendSupportedFillSize(size_t Size) { @@ -1179,7 +1178,7 @@ __SYCL_SUPPRESS_DEPRECATED_POP template using is_same_type = std::is_same, remove_cv_ref_t>; -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template struct ShouldEnableSetArg { static constexpr bool value = std::is_trivially_copyable_v> @@ -1193,7 +1192,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH || is_same_type::value // Stream || detail::check_if_device_copyable_v>; }; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP /// Sets argument for OpenCL interoperability kernels. /// @@ -1229,9 +1228,9 @@ __SYCL_SUPPRESS_DEPRECATED_POP constexpr int AccessTargetMask = 0x7ff; const access::target target = static_cast( type::sizes[NumArgs] & AccessTargetMask); -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH if (target == target::local) { -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP detail::LocalAccessorBaseHost *LocalAccBase = (detail::LocalAccessorBaseHost *)(FieldArg); setLocalAccessorArgHelper(ArgIndex + NumArgs, *LocalAccBase); diff --git a/sycl/include/sycl/image.hpp b/sycl/include/sycl/image.hpp index d3c4aab2706d4..365157283445b 100644 --- a/sycl/include/sycl/image.hpp +++ b/sycl/include/sycl/image.hpp @@ -761,7 +761,7 @@ class unsampled_image #endif public: -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( image_format Format, const range &Range, const property_list &PropList = {}, @@ -776,9 +776,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( image_format Format, const range &Range, AllocatorT Allocator, const property_list &PropList = {}, @@ -794,9 +794,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -814,9 +814,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -836,9 +836,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( void *HostPointer, image_format Format, const range &Range, const property_list &PropList = {}, @@ -853,9 +853,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( void *HostPointer, image_format Format, const range &Range, AllocatorT Allocator, const property_list &PropList = {}, @@ -871,9 +871,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -891,9 +891,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -913,9 +913,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( std::shared_ptr &HostPointer, image_format Format, const range &Range, const property_list &PropList = {}, @@ -930,9 +930,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( std::shared_ptr &HostPointer, image_format Format, const range &Range, AllocatorT Allocator, @@ -949,9 +949,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -970,9 +970,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -992,7 +992,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP #ifdef __DPCPP_ENABLE_UNFINISHED_KHR_EXTENSIONS template &Range, const property_list &PropList = {}, @@ -1214,9 +1214,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format, Sampler); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> sampled_image( @@ -1235,9 +1235,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format, Sampler); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH sampled_image( std::shared_ptr &HostPointer, image_format Format, image_sampler Sampler, const range &Range, @@ -1253,9 +1253,9 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format, Sampler); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> sampled_image( @@ -1274,7 +1274,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format, Sampler); } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP #ifdef __DPCPP_ENABLE_UNFINISHED_KHR_EXTENSIONS template struct LegacyPointerTypes { -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH using decorated_type = typename detail::DecoratedType< ElementType, access::address_space::constant_space>::type; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP using pointer_t = decorated_type *; using const_pointer_t = decorated_type const *; }; @@ -62,10 +62,10 @@ struct LegacyReferenceTypes { template struct LegacyReferenceTypes { -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH using decorated_type = typename detail::DecoratedType< ElementType, access::address_space::constant_space>::type; -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP using reference_t = decorated_type &; using const_reference_t = decorated_type &; }; @@ -262,7 +262,7 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { m_Pointer = nullptr; return *this; } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template < access::address_space OtherSpace, access::decorated OtherIsDecorated, typename = @@ -273,8 +273,8 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH m_Pointer = detail::static_address_cast(Other.get_decorated()); return *this; } -__SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_PUSH template < access::address_space OtherSpace, access::decorated OtherIsDecorated, typename = @@ -285,7 +285,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH m_Pointer = detail::static_address_cast(std::move(Other.m_Pointer)); return *this; } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP reference operator*() const { return *m_Pointer; } pointer operator->() const { return get(); } @@ -571,11 +571,11 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { // Legacy has a different interface. static_assert(DecorateAddress != access::decorated::legacy); // constant_space is only supported in legacy multi_ptr. -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH static_assert(Space != access::address_space::constant_space, "SYCL 2020 multi_ptr does not support the deprecated " "constant_space address space."); -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP // Constructors multi_ptr() : m_Pointer(nullptr) {} @@ -823,7 +823,7 @@ class multi_ptr { Accessor.get_pointer().get())) {} // Only if Space == local_space || generic_space -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template < int dimensions, access::mode Mode, access::placeholder isPlaceholder, typename PropertyListT, access::address_space _Space = Space, @@ -834,7 +834,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH isPlaceholder, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == local_space || generic_space template @@ -842,7 +842,7 @@ __SYCL_SUPPRESS_DEPRECATED_POP : multi_ptr(Accessor.get_pointer()) {} // Only if Space == constant_space -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template < int dimensions, access::mode Mode, access::placeholder isPlaceholder, typename PropertyListT, access::address_space _Space = Space, @@ -852,7 +852,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH isPlaceholder, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP // The following constructors are necessary to create multi_ptr from @@ -881,7 +881,7 @@ __SYCL_SUPPRESS_DEPRECATED_POP : multi_ptr(Accessor.get_pointer()) {} // Only if Space == local_space || generic_space and element type is const -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template Accessor) : multi_ptr(Accessor.get_pointer()) {} -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == local_space || generic_space and element type is const template Accessor) : multi_ptr(Accessor.get_pointer()) {} -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP // TODO: This constructor is the temporary solution for the existing problem // with conversions from multi_ptr { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == local_space || generic_space -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::mode Mode, typename PropertyListT, access::address_space _Space = Space, @@ -1139,10 +1139,10 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH access::placeholder::false_t, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == local_space || generic_space -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::address_space _Space = Space, @@ -1151,10 +1151,10 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH Space == access::address_space::local_space)>> multi_ptr(local_accessor Accessor) : multi_ptr(Accessor.get_pointer()) {} -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == constant_space -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::mode Mode, typename PropertyListT, access::address_space _Space = Space, @@ -1164,7 +1164,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH access::placeholder::false_t, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP using ReturnPtr = detail::const_if_const_AS *; // Returns the underlying OpenCL C pointer @@ -1293,7 +1293,7 @@ class multi_ptr { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == local_space || generic_space -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::mode Mode, typename PropertyListT, access::address_space _Space = Space, @@ -1304,10 +1304,10 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH access::placeholder::false_t, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == local_space || generic_space -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::address_space _Space = Space, @@ -1316,10 +1316,10 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH Space == access::address_space::local_space)>> multi_ptr(local_accessor Accessor) : multi_ptr(Accessor.get_pointer()) {} -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == constant_space -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::mode Mode, typename PropertyListT, access::address_space _Space = Space, @@ -1329,7 +1329,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH access::placeholder::false_t, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP // Returns the underlying OpenCL C pointer pointer_t get() const { return m_Pointer; } diff --git a/sycl/include/sycl/nd_item.hpp b/sycl/include/sycl/nd_item.hpp index 052492ab9c647..82c25dedec964 100644 --- a/sycl/include/sycl/nd_item.hpp +++ b/sycl/include/sycl/nd_item.hpp @@ -71,9 +71,9 @@ template class nd_item { size_t LinId = 0; id Index = get_global_id(); range Extent = get_global_range(); -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH id Offset = get_offset(); -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP if (1 == Dimensions) { LinId = Index[0] - Offset[0]; } else if (2 == Dimensions) { @@ -200,18 +200,18 @@ __SYCL_SUPPRESS_DEPRECATED_POP } nd_range get_nd_range() const noexcept { -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH return nd_range(get_global_range(), get_local_range(), get_offset()); -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP } #ifndef __INTEL_PREVIEW_BREAKING_CHANGES -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH __SYCL2020_DEPRECATED("use sycl::group_barrier() free function instead") void barrier([[maybe_unused]] access::fence_space accessSpace = access::fence_space::global_and_local) const { -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP #ifdef __SYCL_DEVICE_ONLY__ uint32_t flags = _V1::detail::getSPIRVMemorySemanticsMask(accessSpace); __spirv_ControlBarrier(__spv::Scope::Workgroup, __spv::Scope::Workgroup, @@ -221,7 +221,7 @@ __SYCL_SUPPRESS_DEPRECATED_POP /// Executes a work-group mem-fence with memory ordering on the local address /// space, global address space or both based on the value of \p accessSpace. -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH template __SYCL2020_DEPRECATED("use sycl::atomic_fence() free function instead") void mem_fence( @@ -231,7 +231,7 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH accessMode == access::mode::read_write, access::fence_space> accessSpace = access::fence_space::global_and_local) const { -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP #if __SYCL_DEVICE_ONLY__ uint32_t flags = detail::getSPIRVMemorySemanticsMask(accessSpace); // TODO: currently, there is no good way in SPIR-V to set the memory diff --git a/sycl/include/sycl/sycl_span.hpp b/sycl/include/sycl/sycl_span.hpp index 23ca7b4722ce8..b72925bd8e720 100644 --- a/sycl/include/sycl/sycl_span.hpp +++ b/sycl/include/sycl/sycl_span.hpp @@ -393,21 +393,21 @@ template class _SYCL_SPAN_TEMPLATE_VIS span { return rev_iterator(begin()); } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH _SYCL_SPAN_INLINE_VISIBILITY span __as_bytes() const noexcept { return span{ reinterpret_cast(data()), size_bytes()}; } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH _SYCL_SPAN_INLINE_VISIBILITY span __as_writable_bytes() const noexcept { return span{ reinterpret_cast(data()), size_bytes()}; } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP private: pointer __data; @@ -590,19 +590,19 @@ class _SYCL_SPAN_TEMPLATE_VIS span<_Tp, dynamic_extent> { return rev_iterator(begin()); } -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH _SYCL_SPAN_INLINE_VISIBILITY span __as_bytes() const noexcept { return {reinterpret_cast(data()), size_bytes()}; } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL_SUPPRESS_DEPRECATED_PUSH _SYCL_SPAN_INLINE_VISIBILITY span __as_writable_bytes() const noexcept { return {reinterpret_cast(data()), size_bytes()}; } -__SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_POP private: pointer __data; From e905a004fddfaad0cd9b70446e16d58ff26f3530 Mon Sep 17 00:00:00 2001 From: Jakub Stepnowski Date: Thu, 22 Oct 2026 02:29:30 +0200 Subject: [PATCH 03/11] [SYCL] Suppress internal deprecation warnings in headers Suppress internal deprecation warnings in SYCL headers when the include directory is specified with -I, as discussed in: https://github.com/intel/llvm/issues/22038. --- sycl/include/sycl/access/access.hpp | 2 ++ sycl/include/sycl/accessor.hpp | 13 ++++++------- sycl/include/sycl/detail/address_space_cast.hpp | 1 + sycl/include/sycl/detail/helpers.hpp | 4 +--- sycl/include/sycl/group.hpp | 2 +- sycl/include/sycl/handler.hpp | 10 +++++++--- sycl/include/sycl/multi_ptr.hpp | 10 ++++++++++ sycl/include/sycl/nd_item.hpp | 8 +++++--- 8 files changed, 33 insertions(+), 17 deletions(-) diff --git a/sycl/include/sycl/access/access.hpp b/sycl/include/sycl/access/access.hpp index a1436e8223527..52bfa4ef42196 100644 --- a/sycl/include/sycl/access/access.hpp +++ b/sycl/include/sycl/access/access.hpp @@ -106,11 +106,13 @@ template struct mode_target_tag_t { inline constexpr mode_tag_t read_only{}; inline constexpr mode_tag_t read_write{}; inline constexpr mode_tag_t write_only{}; + __SYCL_SUPPRESS_DEPRECATED_PUSH __SYCL2020_DEPRECATED("use 'read_only' instead") inline constexpr mode_target_tag_t read_constant{}; __SYCL_SUPPRESS_DEPRECATED_POP + inline constexpr mode_target_tag_t read_only_host_task; inline constexpr mode_target_tag_t diff --git a/sycl/include/sycl/accessor.hpp b/sycl/include/sycl/accessor.hpp index e88e66280a5f2..df40549073ca9 100644 --- a/sycl/include/sycl/accessor.hpp +++ b/sycl/include/sycl/accessor.hpp @@ -293,6 +293,7 @@ class accessor_common { __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsHostBuf = AccessTarget == access::target::host_buffer; __SYCL_SUPPRESS_DEPRECATED_POP + constexpr static bool IsHostTask = AccessTarget == access::target::host_task; // SYCL2020 4.7.6.9.4.3 // IsPlaceHolder template parameter has no bearing on whether the accessor @@ -455,7 +456,6 @@ constexpr access::target deduceAccessTarget(access::target defaultTarget) { access::target::constant_buffer>>) { return access::target::constant_buffer; } - __SYCL_SUPPRESS_DEPRECATED_POP if constexpr ( std::is_same_v class LocalAccessorBaseDevice { public: @@ -2099,11 +2100,9 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : #endif public detail::accessor_common { - __SYCL_SUPPRESS_DEPRECATED_POP protected: constexpr static int AdjustedDim = Dimensions == 0 ? 1 : Dimensions; - __SYCL_SUPPRESS_DEPRECATED_PUSH using AccessorCommonT = detail::accessor_common; @@ -2360,8 +2359,8 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : } }; -__SYCL_SUPPRESS_DEPRECATED_PUSH // TODO: Remove deprecated specialization once no longer needed +__SYCL_SUPPRESS_DEPRECATED_PUSH template class __SYCL_EBO __SYCL_SPECIAL_CLASS accessor< @@ -2621,9 +2620,9 @@ class __SYCL_EBO host_accessor access::placeholder::false_t>::ConcreteASPtrType Ptr, range AccessRange, range MemRange, id Offset) { - __SYCL_SUPPRESS_DEPRECATED_POP AccessorT::__init(Ptr, AccessRange, MemRange, Offset); } + __SYCL_SUPPRESS_DEPRECATED_POP #ifndef __SYCL_DEVICE_ONLY__ host_accessor(const detail::AccessorImplPtr &Impl) @@ -2778,20 +2777,20 @@ class __SYCL_EBO host_accessor !IsAccessReadOnly && Dims == 0>> const host_accessor & operator=(const typename AccessorT::value_type &Other) const { - __SYCL_SUPPRESS_DEPRECATED_POP *AccessorT::getQualifiedPtr() = Other; return *this; } + __SYCL_SUPPRESS_DEPRECATED_POP __SYCL_SUPPRESS_DEPRECATED_PUSH template > const host_accessor &operator=(typename AccessorT::value_type &&Other) const { - __SYCL_SUPPRESS_DEPRECATED_POP *AccessorT::getQualifiedPtr() = std::move(Other); return *this; } + __SYCL_SUPPRESS_DEPRECATED_POP // implicit conversion between const / non-const types for read only accessors template ( __spv::MemorySemanticsMask::SequentiallyConsistent | @@ -232,8 +230,8 @@ getSPIRVMemorySemanticsMask(const access::fence_space AccessSpace, __spv::MemorySemanticsMask::SequentiallyConsistent | __spv::MemorySemanticsMask::CrossWorkgroupMemory | LocalScopeMask); - __SYCL_SUPPRESS_DEPRECATED_POP } +__SYCL_SUPPRESS_DEPRECATED_POP #endif // __INTEL_PREVIEW_BREAKING_CHANGES inline constexpr bool is_power_of_two(int x) { return (x & (x - 1)) == 0; } diff --git a/sycl/include/sycl/group.hpp b/sycl/include/sycl/group.hpp index e138b27455fb3..77440cd1a33ac 100644 --- a/sycl/include/sycl/group.hpp +++ b/sycl/include/sycl/group.hpp @@ -319,7 +319,6 @@ template class __SYCL_TYPE(group) group { accessMode == access::mode::read_write, access::fence_space> accessSpace = access::fence_space::global_and_local) const { - __SYCL_SUPPRESS_DEPRECATED_POP #ifdef __SYCL_DEVICE_ONLY__ uint32_t flags = detail::getSPIRVMemorySemanticsMask(accessSpace); // TODO: currently, there is no good way in SPIR-V to set the memory @@ -333,6 +332,7 @@ template class __SYCL_TYPE(group) group { __spirv_MemoryBarrier(__spv::Scope::Workgroup, flags); #endif } + __SYCL_SUPPRESS_DEPRECATED_POP #endif // __INTEL_PREVIEW_BREAKING_CHANGES /// Asynchronously copies a number of elements specified by \p numElements diff --git a/sycl/include/sycl/handler.hpp b/sycl/include/sycl/handler.hpp index 31e5b03c462a3..485dfc2cf545f 100644 --- a/sycl/include/sycl/handler.hpp +++ b/sycl/include/sycl/handler.hpp @@ -492,12 +492,13 @@ class __SYCL_EXPORT handler { void setArgHelper(int ArgIndex, accessor &&Arg) { - __SYCL_SUPPRESS_DEPRECATED_POP(void) ArgIndex; + (void)ArgIndex; (void)Arg; #ifndef __SYCL_DEVICE_ONLY__ setLocalAccessorArgHelper(ArgIndex, Arg); #endif } + __SYCL_SUPPRESS_DEPRECATED_POP // setArgHelper for local accessor argument (up to date accessor interface) template @@ -519,7 +520,6 @@ class __SYCL_EXPORT handler { setArgHelper( int ArgIndex, accessor &&Arg) { - __SYCL_SUPPRESS_DEPRECATED_POP detail::AccessorBaseHost *AccBase = (detail::AccessorBaseHost *)&Arg; const detail::AccessorImplPtr &AccImpl = detail::getSyclObjImpl(*AccBase); // Ensure the data of AccImpl lives at least as long as the handler. @@ -529,6 +529,7 @@ class __SYCL_EXPORT handler { addArg(detail::kernel_param_kind_t::kind_accessor, Req, static_cast(AccessTarget), ArgIndex); } + __SYCL_SUPPRESS_DEPRECATED_POP template void setArgHelper(int ArgIndex, T &&Arg) { void *StoredArg = storePlainArg(Arg); @@ -1228,12 +1229,15 @@ class __SYCL_EXPORT handler { constexpr int AccessTargetMask = 0x7ff; const access::target target = static_cast( type::sizes[NumArgs] & AccessTargetMask); + __SYCL_SUPPRESS_DEPRECATED_PUSH + if (target == target::local) { - __SYCL_SUPPRESS_DEPRECATED_POP detail::LocalAccessorBaseHost *LocalAccBase = (detail::LocalAccessorBaseHost *)(FieldArg); setLocalAccessorArgHelper(ArgIndex + NumArgs, *LocalAccBase); + + __SYCL_SUPPRESS_DEPRECATED_POP } else { detail::AccessorBaseHost *AccBase = (detail::AccessorBaseHost *)(FieldArg); diff --git a/sycl/include/sycl/multi_ptr.hpp b/sycl/include/sycl/multi_ptr.hpp index a1d278e4c9358..086a04472dcd0 100644 --- a/sycl/include/sycl/multi_ptr.hpp +++ b/sycl/include/sycl/multi_ptr.hpp @@ -39,10 +39,12 @@ struct LegacyPointerTypes { // with the unsupported space. template struct LegacyPointerTypes { + __SYCL_SUPPRESS_DEPRECATED_PUSH using decorated_type = typename detail::DecoratedType< ElementType, access::address_space::constant_space>::type; __SYCL_SUPPRESS_DEPRECATED_POP + using pointer_t = decorated_type *; using const_pointer_t = decorated_type const *; }; @@ -62,10 +64,12 @@ struct LegacyReferenceTypes { template struct LegacyReferenceTypes { + __SYCL_SUPPRESS_DEPRECATED_PUSH using decorated_type = typename detail::DecoratedType< ElementType, access::address_space::constant_space>::type; __SYCL_SUPPRESS_DEPRECATED_POP + using reference_t = decorated_type &; using const_reference_t = decorated_type &; }; @@ -262,6 +266,7 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { m_Pointer = nullptr; return *this; } + __SYCL_SUPPRESS_DEPRECATED_PUSH template < access::address_space OtherSpace, access::decorated OtherIsDecorated, @@ -274,6 +279,7 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { return *this; } __SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_PUSH template < access::address_space OtherSpace, access::decorated OtherIsDecorated, @@ -570,6 +576,7 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { std::add_pointer_t>); // Legacy has a different interface. static_assert(DecorateAddress != access::decorated::legacy); + // constant_space is only supported in legacy multi_ptr. __SYCL_SUPPRESS_DEPRECATED_PUSH static_assert(Space != access::address_space::constant_space, @@ -1378,6 +1385,7 @@ template ) -> multi_ptr; + __SYCL_SUPPRESS_DEPRECATED_PUSH template @@ -1386,6 +1394,7 @@ multi_ptr(accessor multi_ptr; __SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_PUSH template @@ -1393,6 +1402,7 @@ multi_ptr( accessor) -> multi_ptr; __SYCL_SUPPRESS_DEPRECATED_POP + template multi_ptr(local_accessor) -> multi_ptr; diff --git a/sycl/include/sycl/nd_item.hpp b/sycl/include/sycl/nd_item.hpp index 82c25dedec964..7ec534149714a 100644 --- a/sycl/include/sycl/nd_item.hpp +++ b/sycl/include/sycl/nd_item.hpp @@ -71,9 +71,11 @@ template class nd_item { size_t LinId = 0; id Index = get_global_id(); range Extent = get_global_range(); + __SYCL_SUPPRESS_DEPRECATED_PUSH id Offset = get_offset(); __SYCL_SUPPRESS_DEPRECATED_POP + if (1 == Dimensions) { LinId = Index[0] - Offset[0]; } else if (2 == Dimensions) { @@ -206,18 +208,18 @@ template class nd_item { __SYCL_SUPPRESS_DEPRECATED_POP } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES __SYCL_SUPPRESS_DEPRECATED_PUSH +#ifndef __INTEL_PREVIEW_BREAKING_CHANGES __SYCL2020_DEPRECATED("use sycl::group_barrier() free function instead") void barrier([[maybe_unused]] access::fence_space accessSpace = access::fence_space::global_and_local) const { - __SYCL_SUPPRESS_DEPRECATED_POP #ifdef __SYCL_DEVICE_ONLY__ uint32_t flags = _V1::detail::getSPIRVMemorySemanticsMask(accessSpace); __spirv_ControlBarrier(__spv::Scope::Workgroup, __spv::Scope::Workgroup, flags); #endif } + __SYCL_SUPPRESS_DEPRECATED_POP /// Executes a work-group mem-fence with memory ordering on the local address /// space, global address space or both based on the value of \p accessSpace. @@ -231,7 +233,6 @@ template class nd_item { accessMode == access::mode::read_write, access::fence_space> accessSpace = access::fence_space::global_and_local) const { - __SYCL_SUPPRESS_DEPRECATED_POP #if __SYCL_DEVICE_ONLY__ uint32_t flags = detail::getSPIRVMemorySemanticsMask(accessSpace); // TODO: currently, there is no good way in SPIR-V to set the memory @@ -245,6 +246,7 @@ template class nd_item { __spirv_MemoryBarrier(__spv::Scope::Workgroup, flags); #endif } + __SYCL_SUPPRESS_DEPRECATED_POP #endif // __INTEL_PREVIEW_BREAKING_CHANGES /// Asynchronously copies a number of elements specified by \p numElements From 7f846c3b88646b31fb2fbe332d5cb0bd489c53fc Mon Sep 17 00:00:00 2001 From: Jakub Stepnowski Date: Fri, 23 Oct 2026 00:56:44 +0200 Subject: [PATCH 04/11] [SYCL] Suppress internal deprecation warnings in headers Suppress internal deprecation warnings in SYCL headers when the include directory is specified with -I, as discussed in: #22038. --- sycl/include/sycl/access/access.hpp | 2 +- sycl/include/sycl/accessor_image.hpp | 58 +++++++++++++++++++ sycl/include/sycl/aliases.hpp | 2 +- sycl/include/sycl/backend_types.hpp | 2 +- .../sycl/detail/address_space_cast.hpp | 1 + sycl/include/sycl/detail/helpers.hpp | 7 ++- sycl/include/sycl/detail/image_ocl_types.hpp | 3 +- .../sycl/ext/oneapi/experimental/fragment.hpp | 7 +++ .../sycl/ext/oneapi/filter_selector.hpp | 12 +++- sycl/include/sycl/group.hpp | 2 +- sycl/include/sycl/handler.hpp | 2 +- sycl/include/sycl/interop_handle.hpp | 17 ++++-- sycl/include/sycl/multi_ptr.hpp | 2 +- sycl/include/sycl/nd_item.hpp | 6 +- sycl/include/sycl/reduction.hpp | 43 ++++++++------ sycl/include/sycl/sampler.hpp | 3 +- sycl/include/sycl/stream.hpp | 10 +++- sycl/include/sycl/sycl_span.hpp | 11 ++-- .../Basic/accessor/Inputs/device_accessor.cpp | 1 + 19 files changed, 144 insertions(+), 47 deletions(-) diff --git a/sycl/include/sycl/access/access.hpp b/sycl/include/sycl/access/access.hpp index 52bfa4ef42196..cc3023e2ae482 100644 --- a/sycl/include/sycl/access/access.hpp +++ b/sycl/include/sycl/access/access.hpp @@ -8,7 +8,7 @@ #pragma once -#include // for __SYCL2020_DEPRECATED +#include // for __SYCL2020_DEPRECATED, __SY... #ifdef __SYCL_DEVICE_ONLY__ #include diff --git a/sycl/include/sycl/accessor_image.hpp b/sycl/include/sycl/accessor_image.hpp index be5eb3a697ae4..0bb0519a38761 100644 --- a/sycl/include/sycl/accessor_image.hpp +++ b/sycl/include/sycl/accessor_image.hpp @@ -1,6 +1,7 @@ #pragma once #include +#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP #include #include #include @@ -220,18 +221,34 @@ class image_accessor template friend class __image_array_slice__; + __SYCL_SUPPRESS_DEPRECATED_PUSH + constexpr static bool IsHostImageAcc = (AccessTarget == access::target::host_image); + __SYCL_SUPPRESS_DEPRECATED_POP + + __SYCL_SUPPRESS_DEPRECATED_PUSH + constexpr static bool IsImageAcc = (AccessTarget == access::target::image); + __SYCL_SUPPRESS_DEPRECATED_POP + + __SYCL_SUPPRESS_DEPRECATED_PUSH + constexpr static bool IsImageArrayAcc = (AccessTarget == access::target::image_array); + __SYCL_SUPPRESS_DEPRECATED_POP + + __SYCL_SUPPRESS_DEPRECATED_PUSH + constexpr static bool IsImageAccessWriteOnly = (AccessMode == access::mode::write || AccessMode == access::mode::discard_write); + __SYCL_SUPPRESS_DEPRECATED_POP + constexpr static bool IsImageAccessAnyWrite = (IsImageAccessWriteOnly || AccessMode == access::mode::read_write); @@ -433,6 +450,8 @@ class image_accessor #endif + __SYCL_SUPPRESS_DEPRECATED_PUSH + // Available only when: // (accessTarget == access::target::image && accessMode == access::mode::read) // || (accessTarget == access::target::host_image && ( accessMode == @@ -453,6 +472,10 @@ class image_accessor #endif } + __SYCL_SUPPRESS_DEPRECATED_POP + + __SYCL_SUPPRESS_DEPRECATED_PUSH + // Available only when: // (accessTarget == access::target::image && accessMode == access::mode::read) // || (accessTarget == access::target::host_image && ( accessMode == @@ -475,6 +498,8 @@ class image_accessor #endif } + __SYCL_SUPPRESS_DEPRECATED_POP + // Available only when: // (accessTarget == access::target::image && (accessMode == // access::mode::write || accessMode == access::mode::discard_write)) || @@ -528,6 +553,8 @@ class __image_array_slice__ { } public: + __SYCL_SUPPRESS_DEPRECATED_PUSH + __image_array_slice__( accessor> @@ -535,6 +562,8 @@ class __image_array_slice__ { size_t Idx) : MBaseAcc(BaseAcc), MIdx(Idx) {} + __SYCL_SUPPRESS_DEPRECATED_POP + template < typename CoordT, int Dims = Dimensions, typename = std::enable_if_t<(IsValidCoordDataT::value)>> @@ -542,12 +571,16 @@ class __image_array_slice__ { return MBaseAcc.read(getAdjustedCoords(Coords)); } + __SYCL_SUPPRESS_DEPRECATED_PUSH + template ::value>> DataT read(const CoordT &Coords, const sampler &Smpl) const { return MBaseAcc.read(getAdjustedCoords(Coords), Smpl); } + __SYCL_SUPPRESS_DEPRECATED_POP + template ::value>> void write(const CoordT &Coords, const DataT &Color) const { @@ -588,12 +621,20 @@ class __image_array_slice__ { private: size_t MIdx; + + __SYCL_SUPPRESS_DEPRECATED_PUSH + accessor> MBaseAcc; + + __SYCL_SUPPRESS_DEPRECATED_POP }; } // namespace detail + +__SYCL_SUPPRESS_DEPRECATED_PUSH + /// Image accessors. /// /// Available only when accessTarget == access::target::image. @@ -626,6 +667,10 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor< #endif } + __SYCL_SUPPRESS_DEPRECATED_POP + + __SYCL_SUPPRESS_DEPRECATED_PUSH + template accessor(sycl::image &Image, handler &CommandGroupHandler, const property_list &propList) @@ -657,6 +702,10 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor< #endif }; +__SYCL_SUPPRESS_DEPRECATED_POP + +__SYCL_SUPPRESS_DEPRECATED_PUSH + /// Host image accessor. /// /// Available only when accessTarget == access::target::host_image. @@ -690,6 +739,10 @@ class __SYCL_EBO accessor accessor(sycl::image &Image, handler &CommandGroupHandler, const property_list &propList) @@ -757,6 +814,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor< } }; +__SYCL_SUPPRESS_DEPRECATED_POP // SYCL 2020 image accessors template // for __SYCL2020_DEPRECATED +#include // for __SYCL2020_DEPRECATED, __SY... #include // for uint8_t, int16_t, int32_t diff --git a/sycl/include/sycl/backend_types.hpp b/sycl/include/sycl/backend_types.hpp index 50fea382a8d5d..7ef3c24909408 100644 --- a/sycl/include/sycl/backend_types.hpp +++ b/sycl/include/sycl/backend_types.hpp @@ -8,7 +8,7 @@ #pragma once -#include // for __SYCL2020_DEPRECATED +#include // for __SYCL2020_DEPRECATED, __SY... #include // for __SYCL_EXPORT #include // for ostream diff --git a/sycl/include/sycl/detail/address_space_cast.hpp b/sycl/include/sycl/detail/address_space_cast.hpp index 291ae97738b82..13a1d2f9e2dd1 100644 --- a/sycl/include/sycl/detail/address_space_cast.hpp +++ b/sycl/include/sycl/detail/address_space_cast.hpp @@ -10,6 +10,7 @@ #include #include +#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP #include diff --git a/sycl/include/sycl/detail/helpers.hpp b/sycl/include/sycl/detail/helpers.hpp index ac7e9658f808a..adc30b71038f5 100644 --- a/sycl/include/sycl/detail/helpers.hpp +++ b/sycl/include/sycl/detail/helpers.hpp @@ -9,9 +9,10 @@ #pragma once #include // for MemorySemanticsMask -#include // for fence_space -#include // for __SYCL_EXPORT -#include // for memory_order +#include // for fence_space +#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP +#include // for __SYCL_EXPORT +#include // for memory_order #ifdef __SYCL_DEVICE_ONLY__ #include diff --git a/sycl/include/sycl/detail/image_ocl_types.hpp b/sycl/include/sycl/detail/image_ocl_types.hpp index 6524278ecf631..429b6275f50cc 100644 --- a/sycl/include/sycl/detail/image_ocl_types.hpp +++ b/sycl/include/sycl/detail/image_ocl_types.hpp @@ -32,6 +32,7 @@ #include #include +#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP namespace sycl { inline namespace _V1 { @@ -561,7 +562,7 @@ struct opencl_image_type { using type = opencl_image_type *; }; -__SYCL_SUPPRESS_DEPRECATED_PUSH +__SYCL_SUPPRESS_DEPRECATED_POP template struct sampled_opencl_image_type { using type = void *; diff --git a/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp b/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp index ae6b4b531960d..8d6514f37b752 100644 --- a/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp +++ b/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp @@ -9,6 +9,7 @@ #pragma once #include +#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP #include #include #include @@ -185,7 +186,13 @@ __SYCL_ALWAYS_INLINE inline fragment get_opportunistic_group() { #ifdef __SYCL_DEVICE_ONLY__ #if defined(__SPIR__) || defined(__SPIRV__) + + __SYCL_SUPPRESS_DEPRECATED_PUSH + sycl::sub_group sg = sycl::ext::oneapi::experimental::this_sub_group(); + + __SYCL_SUPPRESS_DEPRECATED_POP + sub_group_mask mask = sycl::ext::oneapi::group_ballot(sg, true); return fragment(mask, 0, 1); #elif defined(__NVPTX__) diff --git a/sycl/include/sycl/ext/oneapi/filter_selector.hpp b/sycl/include/sycl/ext/oneapi/filter_selector.hpp index ab5b33324fd30..3948b08497a90 100644 --- a/sycl/include/sycl/ext/oneapi/filter_selector.hpp +++ b/sycl/include/sycl/ext/oneapi/filter_selector.hpp @@ -8,9 +8,10 @@ #pragma once -#include // for __SYCL_EXPORT -#include // for device -#include // for device_selector +#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP +#include // for __SYCL_EXPORT +#include // for device +#include // for device_selector #include // for shared_ptr #include // for string @@ -34,6 +35,8 @@ namespace detail { class filter_selector_impl; } // namespace detail +__SYCL_SUPPRESS_DEPRECATED_PUSH + class __SYCL_EXPORT_DEPRECATED( "filter_selector is deprecated, use a SYCL 2020 device selector callable " "instead") filter_selector : public device_selector { @@ -51,6 +54,9 @@ class __SYCL_EXPORT_DEPRECATED( std::shared_ptr impl; filter_selector(sycl::detail::string_view filter); }; + +__SYCL_SUPPRESS_DEPRECATED_POP + } // namespace ext::oneapi #ifdef __SYCL_INTERNAL_API diff --git a/sycl/include/sycl/group.hpp b/sycl/include/sycl/group.hpp index 77440cd1a33ac..913843b3fb18c 100644 --- a/sycl/include/sycl/group.hpp +++ b/sycl/include/sycl/group.hpp @@ -13,7 +13,7 @@ #include #include // for convertToOpenCLGroupAsyncCopyPtr #include // for __SYCL_TYPE -#include // for __SYCL2020_DEPRECATED +#include // for __SYCL2020_DEPRECATED, __SY... #include #include // for Builder, getSPIRVMemo... #include // for NDLoop diff --git a/sycl/include/sycl/handler.hpp b/sycl/include/sycl/handler.hpp index 485dfc2cf545f..14ec3cb565737 100644 --- a/sycl/include/sycl/handler.hpp +++ b/sycl/include/sycl/handler.hpp @@ -12,7 +12,7 @@ #include #include #include -#include +#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP #include #include #include diff --git a/sycl/include/sycl/interop_handle.hpp b/sycl/include/sycl/interop_handle.hpp index 9a521345a708d..cd09f1b3aa223 100644 --- a/sycl/include/sycl/interop_handle.hpp +++ b/sycl/include/sycl/interop_handle.hpp @@ -8,12 +8,13 @@ #pragma once -#include // for target, mode, place... -#include // for AccessorBaseHost -#include // for backend, backend_re... -#include // for buffer -#include // for __SYCL_EXPORT -#include // for getSyclObjImpl +#include // for target, mode, place... +#include // for AccessorBaseHost +#include // for backend, backend_re... +#include // for buffer +#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP +#include // for __SYCL_EXPORT +#include // for getSyclObjImpl #include #include // for accessor_property_list #include // for command_graph @@ -54,6 +55,8 @@ class interop_handle { /// a backend graph object is available for interop. __SYCL_EXPORT bool ext_codeplay_has_graph() const noexcept; + __SYCL_SUPPRESS_DEPRECATED_PUSH + /// Receives a SYCL accessor that has been defined as a requirement for the /// command group, and returns the underlying OpenCL memory object that is /// used by the SYCL runtime. If the accessor passed as parameter is not part @@ -84,6 +87,8 @@ class interop_handle { #endif } + __SYCL_SUPPRESS_DEPRECATED_POP + /// Receives a SYCL accessor that has been defined as a requirement for the /// command group, and returns the underlying OpenCL memory object that is /// used by the SYCL runtime. If the accessor passed as parameter is not part diff --git a/sycl/include/sycl/multi_ptr.hpp b/sycl/include/sycl/multi_ptr.hpp index 086a04472dcd0..dc4623a336f50 100644 --- a/sycl/include/sycl/multi_ptr.hpp +++ b/sycl/include/sycl/multi_ptr.hpp @@ -10,7 +10,7 @@ #include #include -#include +#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP #include #include #include diff --git a/sycl/include/sycl/nd_item.hpp b/sycl/include/sycl/nd_item.hpp index 7ec534149714a..d742163a0ff84 100644 --- a/sycl/include/sycl/nd_item.hpp +++ b/sycl/include/sycl/nd_item.hpp @@ -208,8 +208,9 @@ template class nd_item { __SYCL_SUPPRESS_DEPRECATED_POP } - __SYCL_SUPPRESS_DEPRECATED_PUSH #ifndef __INTEL_PREVIEW_BREAKING_CHANGES + __SYCL_SUPPRESS_DEPRECATED_PUSH + __SYCL2020_DEPRECATED("use sycl::group_barrier() free function instead") void barrier([[maybe_unused]] access::fence_space accessSpace = access::fence_space::global_and_local) const { @@ -219,11 +220,12 @@ template class nd_item { flags); #endif } + __SYCL_SUPPRESS_DEPRECATED_POP + __SYCL_SUPPRESS_DEPRECATED_PUSH /// Executes a work-group mem-fence with memory ordering on the local address /// space, global address space or both based on the value of \p accessSpace. - __SYCL_SUPPRESS_DEPRECATED_PUSH template __SYCL2020_DEPRECATED("use sycl::atomic_fence() free function instead") void mem_fence( diff --git a/sycl/include/sycl/reduction.hpp b/sycl/include/sycl/reduction.hpp index 09cc49822dcea..4448d88d51b35 100644 --- a/sycl/include/sycl/reduction.hpp +++ b/sycl/include/sycl/reduction.hpp @@ -8,25 +8,26 @@ #pragma once -#include // for address_s... -#include // for local_acc... -#include // for aspect -#include // for IsValidAt... -#include // for atomic_ref -#include // for buffer -#include // for min -#include // for __SYCL_EX... -#include // for is_sgenfloat -#include // for createSyc... -#include // for id -#include // for strategy -#include // for make_tuple -#include // for device -#include // for event -#include // for make_erro... -#include // for queue_impl -#include // for workGroup... -#include // for reduce_ov... +#include // for address_s... +#include // for local_acc... +#include // for aspect +#include // for IsValidAt... +#include // for atomic_ref +#include // for buffer +#include // for min +#include // for __SYCL_SU... +#include // for __SYCL_EX... +#include // for is_sgenfloat +#include // for createSyc... +#include // for id +#include // for strategy +#include // for make_tuple +#include // for device +#include // for event +#include // for make_erro... +#include // for queue_impl +#include // for workGroup... +#include // for reduce_ov... #include #include // for handler #include // for getDeline... @@ -1352,6 +1353,8 @@ struct NDRangeReduction< }); }; + __SYCL_SUPPRESS_DEPRECATED_PUSH + auto device = getDeviceFromHandler(CGH); // Integrated/discrete GPUs have different faster path. For discrete GPUs // fast path requires USM device allocations though, so check for that as @@ -1361,6 +1364,8 @@ struct NDRangeReduction< Rest(Redu.getReadWriteAccessorToInitializedGroupsCounter(CGH)); else Rest(Redu.getGroupsCounterAccDiscrete(CGH)); + + __SYCL_SUPPRESS_DEPRECATED_POP } }; diff --git a/sycl/include/sycl/sampler.hpp b/sycl/include/sycl/sampler.hpp index 813473a478b6b..ecfd29a4455a1 100644 --- a/sycl/include/sycl/sampler.hpp +++ b/sycl/include/sycl/sampler.hpp @@ -10,7 +10,8 @@ #include // for mode, placeholder, target #include // for __SYCL_SPECIAL_CLASS, __SYCL_TYPE -#include // for __SYCL_EXPORT +#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP +#include // for __SYCL_EXPORT #include #include // for getSyclObjImpl #include // for property_list diff --git a/sycl/include/sycl/stream.hpp b/sycl/include/sycl/stream.hpp index 5e914af793e42..b937e038b4bfc 100644 --- a/sycl/include/sycl/stream.hpp +++ b/sycl/include/sycl/stream.hpp @@ -15,7 +15,7 @@ #include // for isinf, isnan, signbit #include // for array #include // for __SYCL_SPECIAL_CLASS, __S... -#include // for __SYCL2020_DEPRECATED +#include // for __SYCL2020_DEPRECATED, __SY... #include // for __SYCL_EXPORT #include // for OwnerLessBase #include @@ -97,9 +97,13 @@ using GlobalBufPtrType = typename detail::DecoratedType::type *; constexpr static int GlobalBufDim = 1; +__SYCL_SUPPRESS_DEPRECATED_PUSH + using GlobalOffsetAccessorT = accessor; +__SYCL_SUPPRESS_DEPRECATED_POP + constexpr static access::address_space GlobalOffsetAS = TargetToAS::AS; using GlobalOffsetPtrType = @@ -644,6 +648,8 @@ writeFloatingPoint(GlobalBufAccessorT &GlobalFlushBuf, size_t FlushBufferSize, : 0); } +__SYCL_SUPPRESS_DEPRECATED_PUSH + // Helper method to update offset in the global buffer atomically according to // the provided size of the data in the flush buffer. Return true if offset is // updated and false in case of overflow. @@ -661,6 +667,8 @@ inline bool updateOffset(GlobalOffsetAccessorT &GlobalOffset, return true; } +__SYCL_SUPPRESS_DEPRECATED_POP + inline void flushBuffer(GlobalOffsetAccessorT &GlobalOffset, GlobalBufAccessorT &GlobalBuf, GlobalBufAccessorT &GlobalFlushBuf, unsigned WIOffset) { diff --git a/sycl/include/sycl/sycl_span.hpp b/sycl/include/sycl/sycl_span.hpp index b72925bd8e720..565c2deaf5b7b 100644 --- a/sycl/include/sycl/sycl_span.hpp +++ b/sycl/include/sycl/sycl_span.hpp @@ -124,11 +124,12 @@ template */ -#include // for array -#include // for assert -#include // for size_t, nullptr_t, ptrdiff_t -#include // for SIZE_MAX -#include // for size, data, distance, reverse_iterator +#include // for array +#include // for assert +#include // for size_t, nullptr_t, ptrdiff_t +#include // for SIZE_MAX +#include // for size, data, distance, reverse_iterator +#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP #include // for enable_if_t, enable_if, remove_cv_t, false_type #include // for declval diff --git a/sycl/test-e2e/Basic/accessor/Inputs/device_accessor.cpp b/sycl/test-e2e/Basic/accessor/Inputs/device_accessor.cpp index f50ad95939f8e..844c2dbb8a7e3 100644 --- a/sycl/test-e2e/Basic/accessor/Inputs/device_accessor.cpp +++ b/sycl/test-e2e/Basic/accessor/Inputs/device_accessor.cpp @@ -7,6 +7,7 @@ //===----------------------------------------------------------------------===// #include #include +#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP int main() { // Non-placeholder accessors. From 87259bf7cfc21fc2d685a084b2b319bf31b969a1 Mon Sep 17 00:00:00 2001 From: Jakub Stepnowski Date: Fri, 23 Oct 2026 19:20:13 +0200 Subject: [PATCH 05/11] [SYCL] Suppress internal deprecation warnings in headers Wraps all deprecated features used in header files in: __SYCL_SUPPRESS_DEPRECATED_PUSH/POP and adds comments which features are deprecated inside such block. fixes issue: https://github.com/intel/llvm/issues/22038. --- sycl/include/sycl/access/access.hpp | 23 +++++--- sycl/include/sycl/accessor.hpp | 57 +++++++++---------- sycl/include/sycl/accessor_image.hpp | 42 +++----------- sycl/include/sycl/aliases.hpp | 1 + sycl/include/sycl/backend_types.hpp | 1 + .../sycl/detail/address_space_cast.hpp | 1 + sycl/include/sycl/detail/helpers.hpp | 1 + sycl/include/sycl/detail/image_ocl_types.hpp | 1 + .../sycl/ext/oneapi/experimental/fragment.hpp | 3 +- .../sycl/ext/oneapi/filter_selector.hpp | 2 +- sycl/include/sycl/group.hpp | 1 + sycl/include/sycl/handler.hpp | 20 ++++--- sycl/include/sycl/image.hpp | 31 +--------- sycl/include/sycl/interop_handle.hpp | 2 +- sycl/include/sycl/multi_ptr.hpp | 38 ++++++------- sycl/include/sycl/nd_item.hpp | 9 +-- sycl/include/sycl/reduction.hpp | 4 +- sycl/include/sycl/sampler.hpp | 1 + sycl/include/sycl/stream.hpp | 4 +- sycl/include/sycl/sycl_span.hpp | 6 +- .../Basic/accessor/Inputs/device_accessor.cpp | 4 ++ 21 files changed, 106 insertions(+), 146 deletions(-) diff --git a/sycl/include/sycl/access/access.hpp b/sycl/include/sycl/access/access.hpp index cc3023e2ae482..1ed4cd5a9bad6 100644 --- a/sycl/include/sycl/access/access.hpp +++ b/sycl/include/sycl/access/access.hpp @@ -107,6 +107,7 @@ inline constexpr mode_tag_t read_only{}; inline constexpr mode_tag_t read_write{}; inline constexpr mode_tag_t write_only{}; +// access::target::constant_buffer is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH __SYCL2020_DEPRECATED("use 'read_only' instead") inline constexpr mode_target_tag_t @@ -122,18 +123,18 @@ inline constexpr mode_target_tag_t namespace detail { +// access::target::host_buffer, host_image and access::mode::atomic are +// deprecated. +__SYCL_SUPPRESS_DEPRECATED_PUSH constexpr bool isTargetHostAccess(access::target T) { - __SYCL_SUPPRESS_DEPRECATED_PUSH return T == access::target::host_buffer || T == access::target::host_image; - __SYCL_SUPPRESS_DEPRECATED_POP } constexpr bool modeNeedsOldData(access::mode m) { - __SYCL_SUPPRESS_DEPRECATED_PUSH return m == access::mode::read || m == access::mode::write || m == access::mode::read_write || m == access::mode::atomic; - __SYCL_SUPPRESS_DEPRECATED_POP } +__SYCL_SUPPRESS_DEPRECATED_POP constexpr bool modeWritesNewData(access::mode m) { return m != access::mode::read; @@ -169,12 +170,14 @@ template <> struct TargetToAS { access::address_space::local_space; }; +// access::target::constant_buffer and address_space::constant_space are +// deprecated. +__SYCL_SUPPRESS_DEPRECATED_PUSH template <> struct TargetToAS { - __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static access::address_space AS = access::address_space::constant_space; - __SYCL_SUPPRESS_DEPRECATED_POP }; +__SYCL_SUPPRESS_DEPRECATED_POP template struct DecoratedType; @@ -194,6 +197,8 @@ struct DecoratedType { using type = __OPENCL_GLOBAL_AS__ ElementType; }; +// access::address_space::constant_space is deprecated. +__SYCL_SUPPRESS_DEPRECATED_PUSH template struct DecoratedType { // Current implementation of address spaces handling leads to possibility @@ -207,6 +212,7 @@ struct DecoratedType { using type = __OPENCL_GLOBAL_AS__ ElementType; #endif }; +__SYCL_SUPPRESS_DEPRECATED_POP template struct DecoratedType { @@ -237,12 +243,13 @@ template struct deduce_AS_impl<__OPENCL_LOCAL_AS__ T> { access::address_space::local_space; }; +// access::address_space::constant_space is deprecated. +__SYCL_SUPPRESS_DEPRECATED_PUSH template struct deduce_AS_impl<__OPENCL_CONSTANT_AS__ T> { - __SYCL_SUPPRESS_DEPRECATED_PUSH static constexpr access::address_space value = access::address_space::constant_space; - __SYCL_SUPPRESS_DEPRECATED_POP }; +__SYCL_SUPPRESS_DEPRECATED_POP template struct deduce_AS diff --git a/sycl/include/sycl/accessor.hpp b/sycl/include/sycl/accessor.hpp index df40549073ca9..5b7bc3cc5755c 100644 --- a/sycl/include/sycl/accessor.hpp +++ b/sycl/include/sycl/accessor.hpp @@ -290,6 +290,7 @@ class accessor_common { protected: constexpr static access::address_space AS = TargetToAS::AS; + // access::target::host_buffer is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsHostBuf = AccessTarget == access::target::host_buffer; __SYCL_SUPPRESS_DEPRECATED_POP @@ -307,26 +308,23 @@ class accessor_common { // placeholder either. Otherwise, the accessor is a placeholder. constexpr static bool IsPlaceH = !IsHostBuf; + // access::target::global_buffer, constant_buffer and + // access::mode::discard_write, discard_read_write, atomic are deprecated. + __SYCL_SUPPRESS_DEPRECATED_PUSH // TODO: SYCL 2020 deprecates four of the target enum values // and replaces them with 2 (device and host_task). May want // to change these constexpr. - __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsGlobalBuf = AccessTarget == access::target::global_buffer; - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsConstantBuf = AccessTarget == access::target::constant_buffer; - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsAccessAnyWrite = AccessMode == access::mode::write || AccessMode == access::mode::read_write || AccessMode == access::mode::discard_write || AccessMode == access::mode::discard_read_write; - __SYCL_SUPPRESS_DEPRECATED_POP constexpr static bool IsAccessReadOnly = AccessMode == access::mode::read; static constexpr bool IsConst = std::is_const_v; @@ -334,7 +332,6 @@ class accessor_common { constexpr static bool IsAccessReadWrite = AccessMode == access::mode::read_write; - __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsAccessAtomic = AccessMode == access::mode::atomic; __SYCL_SUPPRESS_DEPRECATED_POP @@ -409,6 +406,7 @@ constexpr access::mode deduceAccessMode() { return access::mode::write; } + // access::target::constant_buffer is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH if constexpr (std::is_same_v< MayBeTag1, @@ -443,9 +441,10 @@ constexpr access::mode deduceAccessMode() { return access::mode::read_write; } -__SYCL_SUPPRESS_DEPRECATED_PUSH template constexpr access::target deduceAccessTarget(access::target defaultTarget) { + // access::target::constant_buffer is deprecated. + __SYCL_SUPPRESS_DEPRECATED_PUSH if constexpr (std::is_same_v< MayBeTag1, mode_target_tag_t>) { return access::target::constant_buffer; } + __SYCL_SUPPRESS_DEPRECATED_POP if constexpr ( std::is_same_v class LocalAccessorBaseDevice { public: @@ -618,15 +617,14 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : friend sycl::detail::ImplUtils; protected: + // access::target::global_buffer, constant_buffer, host_buffer are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH static_assert((AccessTarget == access::target::global_buffer || AccessTarget == access::target::constant_buffer || AccessTarget == access::target::host_buffer || AccessTarget == access::target::host_task), "Expected buffer type"); - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH static_assert((AccessTarget == access::target::global_buffer || AccessTarget == access::target::host_buffer || AccessTarget == access::target::host_task) || @@ -698,6 +696,8 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : static access::mode getAdjustedMode(const PropertyListT &PropertyList) { access::mode AdjustedMode = AccessMode; + // property::noinit and access::mode::discard_write, discard_read_write + // are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH if (PropertyList.template has_property() || PropertyList.template has_property()) { @@ -1687,6 +1687,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : return *(getQualifiedPtr() + LinearIndex); } + // access_mode::atomic is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template > @@ -1715,6 +1714,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : return getQualifiedPtr()[LinearIndex]; } + // access::mode::atomic and sycl::atomic are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template operator typename std::enable_if_t< @@ -1724,9 +1724,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t<(Dims > 0) && AccessMode == access::mode::atomic, atomic> @@ -1735,9 +1733,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t> @@ -1747,11 +1743,14 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : getQualifiedPtr() + LinearIndex)); } __SYCL_SUPPRESS_DEPRECATED_POP + template 1)>> auto operator[](size_t Index) const { return AccessorSubscript(*this, Index); } + // access::target::host_buffer, constant_buffer and constant_ptr are + // deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template get_pointer() const noexcept { return getPointerAdjusted(); } - __SYCL_SUPPRESS_DEPRECATED_POP template < access::target AccessTarget_ = AccessTarget, @@ -1773,7 +1771,6 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : getPointerAdjusted())); } - __SYCL_SUPPRESS_DEPRECATED_PUSH template > @@ -2088,10 +2085,11 @@ accessor(buffer, handler &, Type1, Type2, Type3, access::placeholder::false_t, ext::oneapi::accessor_property_list>; +// access::target::local is deprecated. +__SYCL_SUPPRESS_DEPRECATED_PUSH /// Local accessor /// /// \ingroup sycl_api_acc -__SYCL_SUPPRESS_DEPRECATED_PUSH template class __SYCL_SPECIAL_CLASS local_accessor_base : @@ -2209,6 +2207,8 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : using reference = DataT &; using const_reference = const DataT &; + // access::target::local is deprecated. + __SYCL_SUPPRESS_DEPRECATED_PUSH template > local_accessor_base(handler &, const detail::code_location CodeLoc = detail::code_location::current()) @@ -2279,6 +2279,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : } #endif + __SYCL_SUPPRESS_DEPRECATED_POP size_t get_size() const { return getSize().size() * sizeof(DataT); } __SYCL2020_DEPRECATED("get_count() is deprecated, please use size() instead") @@ -2312,6 +2313,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : return getQualifiedPtr()[Index]; } + // access::mode::atomic and sycl::atomic are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template operator typename std::enable_if_t< @@ -2320,9 +2322,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : return atomic( multi_ptr(getQualifiedPtr())); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t<(Dims > 0) && AccessMode == access::mode::atomic, atomic> @@ -2331,9 +2331,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : return atomic(multi_ptr( getQualifiedPtr() + LinearIndex)); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t> @@ -2360,6 +2358,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base : }; // TODO: Remove deprecated specialization once no longer needed +// access::target::local is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template @@ -2369,6 +2368,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS accessor< public detail::OwnerLessBase< accessor> { + __SYCL_SUPPRESS_DEPRECATED_POP using local_acc = @@ -2593,6 +2593,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(local_accessor) local_accessor friend class ext::oneapi::experimental::dynamic_local_accessor; }; +// access::target::host_buffer is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template @@ -2604,12 +2605,10 @@ class __SYCL_EBO host_accessor protected: using AccessorT = accessor; - __SYCL_SUPPRESS_DEPRECATED_POP constexpr static int AdjustedDim = Dimensions == 0 ? 1 : Dimensions; constexpr static bool IsAccessReadOnly = AccessMode == access::mode::read; - __SYCL_SUPPRESS_DEPRECATED_PUSH template struct IsSameAsBuffer : std::bool_constant && (Dims > 0) && @@ -2622,13 +2621,12 @@ class __SYCL_EBO host_accessor id Offset) { AccessorT::__init(Ptr, AccessRange, MemRange, Offset); } - __SYCL_SUPPRESS_DEPRECATED_POP - #ifndef __SYCL_DEVICE_ONLY__ host_accessor(const detail::AccessorImplPtr &Impl) : accessor{Impl} {} #endif // __SYCL_DEVICE_ONLY__ + __SYCL_SUPPRESS_DEPRECATED_POP public: host_accessor() : AccessorT() {} @@ -2771,6 +2769,7 @@ class __SYCL_EBO host_accessor : host_accessor(BufferRef, CommandGroupHandler, AccessRange, AccessOffset, PropertyList, CodeLoc) {} + // access_mode::atomic is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template > diff --git a/sycl/include/sycl/accessor_image.hpp b/sycl/include/sycl/accessor_image.hpp index 0bb0519a38761..6c6d8081a650c 100644 --- a/sycl/include/sycl/accessor_image.hpp +++ b/sycl/include/sycl/accessor_image.hpp @@ -221,28 +221,18 @@ class image_accessor template friend class __image_array_slice__; + // access::target::host_image, image, image_array and mode::discard_write are + // deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsHostImageAcc = (AccessTarget == access::target::host_image); - __SYCL_SUPPRESS_DEPRECATED_POP - - __SYCL_SUPPRESS_DEPRECATED_PUSH - constexpr static bool IsImageAcc = (AccessTarget == access::target::image); - __SYCL_SUPPRESS_DEPRECATED_POP - - __SYCL_SUPPRESS_DEPRECATED_PUSH - constexpr static bool IsImageArrayAcc = (AccessTarget == access::target::image_array); - __SYCL_SUPPRESS_DEPRECATED_POP - - __SYCL_SUPPRESS_DEPRECATED_PUSH - constexpr static bool IsImageAccessWriteOnly = (AccessMode == access::mode::write || AccessMode == access::mode::discard_write); @@ -450,6 +440,7 @@ class image_accessor #endif + // sycl::sampler is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH // Available only when: @@ -472,10 +463,6 @@ class image_accessor #endif } - __SYCL_SUPPRESS_DEPRECATED_POP - - __SYCL_SUPPRESS_DEPRECATED_PUSH - // Available only when: // (accessTarget == access::target::image && accessMode == access::mode::read) // || (accessTarget == access::target::host_image && ( accessMode == @@ -553,6 +540,7 @@ class __image_array_slice__ { } public: + // access::target::image_array and sycl::sampler are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH __image_array_slice__( @@ -562,8 +550,6 @@ class __image_array_slice__ { size_t Idx) : MBaseAcc(BaseAcc), MIdx(Idx) {} - __SYCL_SUPPRESS_DEPRECATED_POP - template < typename CoordT, int Dims = Dimensions, typename = std::enable_if_t<(IsValidCoordDataT::value)>> @@ -571,8 +557,6 @@ class __image_array_slice__ { return MBaseAcc.read(getAdjustedCoords(Coords)); } - __SYCL_SUPPRESS_DEPRECATED_PUSH - template ::value>> DataT read(const CoordT &Coords, const sampler &Smpl) const { @@ -622,19 +606,19 @@ class __image_array_slice__ { private: size_t MIdx; + // access::target::image_array is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH accessor> MBaseAcc; - __SYCL_SUPPRESS_DEPRECATED_POP }; } // namespace detail +// access::target::image and sycl::image are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH - /// Image accessors. /// /// Available only when accessTarget == access::target::image. @@ -667,10 +651,6 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor< #endif } - __SYCL_SUPPRESS_DEPRECATED_POP - - __SYCL_SUPPRESS_DEPRECATED_PUSH - template accessor(sycl::image &Image, handler &CommandGroupHandler, const property_list &propList) @@ -701,11 +681,10 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor< accessor() = default; #endif }; - __SYCL_SUPPRESS_DEPRECATED_POP +// access::target::host_image and sycl::image are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH - /// Host image accessor. /// /// Available only when accessTarget == access::target::host_image. @@ -738,11 +717,10 @@ class __SYCL_EBO accessor accessor(sycl::image &Image, handler &CommandGroupHandler, const property_list &propList) diff --git a/sycl/include/sycl/aliases.hpp b/sycl/include/sycl/aliases.hpp index 18037b933387d..c4e520382f401 100644 --- a/sycl/include/sycl/aliases.hpp +++ b/sycl/include/sycl/aliases.hpp @@ -137,6 +137,7 @@ using cl_float = float; using cl_double = double; } // namespace opencl +// sycl::cl_* scalar aliases used by the vector alias macros are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH __SYCL_2020_MAKE_VECTOR_ALIASES_FOR_VECTOR_LENGTH(2) __SYCL_2020_MAKE_VECTOR_ALIASES_FOR_VECTOR_LENGTH(3) diff --git a/sycl/include/sycl/backend_types.hpp b/sycl/include/sycl/backend_types.hpp index 7ef3c24909408..446a2b00c4b14 100644 --- a/sycl/include/sycl/backend_types.hpp +++ b/sycl/include/sycl/backend_types.hpp @@ -44,6 +44,7 @@ __SYCL_EXPORT std::ostream &operator<<(std::ostream &Out, backend be); namespace detail { inline std::string_view get_backend_name_no_vendor(backend Backend) { switch (Backend) { + // backend::host is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH case backend::host: return "host"; diff --git a/sycl/include/sycl/detail/address_space_cast.hpp b/sycl/include/sycl/detail/address_space_cast.hpp index 13a1d2f9e2dd1..4a6270cc72853 100644 --- a/sycl/include/sycl/detail/address_space_cast.hpp +++ b/sycl/include/sycl/detail/address_space_cast.hpp @@ -23,6 +23,7 @@ inline constexpr bool address_space_cast_is_possible(access::address_space Src, access::address_space Dst) { // constant_space is unique and is not interchangeable with any other. + // access::address_space::constant_space is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH auto constant_space = access::address_space::constant_space; __SYCL_SUPPRESS_DEPRECATED_POP diff --git a/sycl/include/sycl/detail/helpers.hpp b/sycl/include/sycl/detail/helpers.hpp index adc30b71038f5..b34ae5f2b4f1e 100644 --- a/sycl/include/sycl/detail/helpers.hpp +++ b/sycl/include/sycl/detail/helpers.hpp @@ -187,6 +187,7 @@ getSPIRVMemorySemanticsMask(memory_order) { } #ifndef __INTEL_PREVIEW_BREAKING_CHANGES +// access::fence_space is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH inline constexpr uint32_t getSPIRVMemorySemanticsMask(const access::fence_space AccessSpace, diff --git a/sycl/include/sycl/detail/image_ocl_types.hpp b/sycl/include/sycl/detail/image_ocl_types.hpp index 429b6275f50cc..35367c77585e1 100644 --- a/sycl/include/sycl/detail/image_ocl_types.hpp +++ b/sycl/include/sycl/detail/image_ocl_types.hpp @@ -553,6 +553,7 @@ inline int getSPIRVElementSize(int ImageChannelType, int ImageChannelOrder) { template struct opencl_image_type; +// access::target::host_image is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH // Creation of dummy ocl types for host_image targets. // These dummy ocl types are needed by the compiler parser for the compilation diff --git a/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp b/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp index 8d6514f37b752..83c89e76bf20d 100644 --- a/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp +++ b/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp @@ -187,10 +187,9 @@ get_opportunistic_group() { #ifdef __SYCL_DEVICE_ONLY__ #if defined(__SPIR__) || defined(__SPIRV__) + // ext::oneapi::experimental::this_sub_group() is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH - sycl::sub_group sg = sycl::ext::oneapi::experimental::this_sub_group(); - __SYCL_SUPPRESS_DEPRECATED_POP sub_group_mask mask = sycl::ext::oneapi::group_ballot(sg, true); diff --git a/sycl/include/sycl/ext/oneapi/filter_selector.hpp b/sycl/include/sycl/ext/oneapi/filter_selector.hpp index 3948b08497a90..52913b6b70780 100644 --- a/sycl/include/sycl/ext/oneapi/filter_selector.hpp +++ b/sycl/include/sycl/ext/oneapi/filter_selector.hpp @@ -35,8 +35,8 @@ namespace detail { class filter_selector_impl; } // namespace detail +// sycl::device_selector and ext::oneapi::filter_selector are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH - class __SYCL_EXPORT_DEPRECATED( "filter_selector is deprecated, use a SYCL 2020 device selector callable " "instead") filter_selector : public device_selector { diff --git a/sycl/include/sycl/group.hpp b/sycl/include/sycl/group.hpp index 913843b3fb18c..b28d836b96da1 100644 --- a/sycl/include/sycl/group.hpp +++ b/sycl/include/sycl/group.hpp @@ -307,6 +307,7 @@ template class __SYCL_TYPE(group) group { } #ifndef __INTEL_PREVIEW_BREAKING_CHANGES + // access::fence_space is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH /// Executes a work-group mem-fence with memory ordering on the local address /// space, global address space or both based on the value of \p accessSpace. diff --git a/sycl/include/sycl/handler.hpp b/sycl/include/sycl/handler.hpp index 14ec3cb565737..bc7a6e6afa67b 100644 --- a/sycl/include/sycl/handler.hpp +++ b/sycl/include/sycl/handler.hpp @@ -473,6 +473,7 @@ class __SYCL_EXPORT handler { void setArgsHelper(int) {} + // access::target::local is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH void setLocalAccessorArgHelper(int ArgIndex, detail::LocalAccessorBaseHost &LocalAccBase) { @@ -483,9 +484,7 @@ class __SYCL_EXPORT handler { addArg(detail::kernel_param_kind_t::kind_accessor, Req, static_cast(access::target::local), ArgIndex); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH // setArgHelper for local accessor argument (legacy accessor interface) template @@ -512,8 +511,9 @@ class __SYCL_EXPORT handler { void setArgHelper(int ArgIndex, detail::work_group_memory_impl &Arg); - __SYCL_SUPPRESS_DEPRECATED_PUSH // setArgHelper for non local accessor argument. + // access::target::local is deprecated. + __SYCL_SUPPRESS_DEPRECATED_PUSH template typename std::enable_if_t @@ -547,6 +547,7 @@ class __SYCL_EXPORT handler { } } + // sycl::sampler is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH void setArgHelper(int ArgIndex, sampler &&Arg) { void *StoredArg = storePlainArg(Arg); @@ -771,14 +772,13 @@ class __SYCL_EXPORT handler { return false; } + // access::target::constant_buffer, image and image_array are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool isConstOrGlobal(access::target AccessTarget) { return AccessTarget == access::target::device || AccessTarget == access::target::constant_buffer; } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool isImageOrImageArray(access::target AccessTarget) { return AccessTarget == access::target::image || AccessTarget == access::target::image_array; @@ -795,6 +795,7 @@ class __SYCL_EXPORT handler { AccessMode == access::mode::read_write; } + // access::mode::discard_write and discard_read_write are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool isValidModeForDestinationAccessor(access::mode AccessMode) { @@ -1179,6 +1180,7 @@ class __SYCL_EXPORT handler { template using is_same_type = std::is_same, remove_cv_ref_t>; + // sycl::sampler is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template struct ShouldEnableSetArg { static constexpr bool value = @@ -1230,14 +1232,12 @@ class __SYCL_EXPORT handler { const access::target target = static_cast( type::sizes[NumArgs] & AccessTargetMask); + // access::target::local is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH - if (target == target::local) { detail::LocalAccessorBaseHost *LocalAccBase = (detail::LocalAccessorBaseHost *)(FieldArg); setLocalAccessorArgHelper(ArgIndex + NumArgs, *LocalAccBase); - - __SYCL_SUPPRESS_DEPRECATED_POP } else { detail::AccessorBaseHost *AccBase = (detail::AccessorBaseHost *)(FieldArg); @@ -1247,6 +1247,7 @@ class __SYCL_EXPORT handler { addArg(type::kinds[NumArgs], Req, type::sizes[NumArgs], ArgIndex + NumArgs); } + __SYCL_SUPPRESS_DEPRECATED_POP } else { // for non-accessors, simply call addArg normally. addArg(type::kinds[NumArgs], FieldArg, type::sizes[NumArgs], @@ -2945,6 +2946,8 @@ class __SYCL_EXPORT handler { const detail::kernel_param_kind_t &Kind = ParamDesc.kind; const access::target AccTarget = static_cast(ParamDesc.info & AccessTargetMask); + // access::target::local is deprecated. + __SYCL_SUPPRESS_DEPRECATED_PUSH if ((Kind == detail::kernel_param_kind_t::kind_accessor) && (AccTarget == target::local)) throw sycl::exception( @@ -2952,6 +2955,7 @@ class __SYCL_EXPORT handler { "A local accessor must not be used in a SYCL kernel function " "that is invoked via single_task or via the simple form of " "parallel_for that takes a range parameter."); + __SYCL_SUPPRESS_DEPRECATED_POP if (Kind == detail::kernel_param_kind_t::kind_work_group_memory || Kind == detail::kernel_param_kind_t::kind_dynamic_work_group_memory) throw sycl::exception( diff --git a/sycl/include/sycl/image.hpp b/sycl/include/sycl/image.hpp index 365157283445b..dbe627519947a 100644 --- a/sycl/include/sycl/image.hpp +++ b/sycl/include/sycl/image.hpp @@ -107,6 +107,7 @@ enum class image_format : unsigned int { using byte = unsigned char; +// sycl::byte is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH using image_allocator = detail::aligned_allocator; __SYCL_SUPPRESS_DEPRECATED_POP @@ -761,6 +762,7 @@ class unsampled_image #endif public: + // sycl::byte is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( image_format Format, const range &Range, @@ -776,9 +778,7 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( image_format Format, const range &Range, AllocatorT Allocator, const property_list &PropList = {}, @@ -794,9 +794,7 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -814,9 +812,7 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -836,9 +832,7 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( void *HostPointer, image_format Format, const range &Range, const property_list &PropList = {}, @@ -853,9 +847,7 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( void *HostPointer, image_format Format, const range &Range, AllocatorT Allocator, const property_list &PropList = {}, @@ -871,9 +863,7 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -891,9 +881,7 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -913,9 +901,7 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( std::shared_ptr &HostPointer, image_format Format, const range &Range, const property_list &PropList = {}, @@ -930,9 +916,7 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH unsampled_image( std::shared_ptr &HostPointer, image_format Format, const range &Range, AllocatorT Allocator, @@ -949,9 +933,7 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -970,9 +952,7 @@ class unsampled_image CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> unsampled_image( @@ -1199,6 +1179,7 @@ class sampled_image #endif public: + // sycl::byte is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH sampled_image( const void *HostPointer, image_format Format, image_sampler Sampler, @@ -1214,9 +1195,7 @@ class sampled_image CodeLoc, (void *)this->impl.get(), nullptr, Dimensions, detail::rangeToArray(Range).data(), Format, Sampler); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> sampled_image( @@ -1235,9 +1214,7 @@ class sampled_image CodeLoc, (void *)this->impl.get(), HostPointer, Dimensions, detail::rangeToArray(Range).data(), Format, Sampler); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH sampled_image( std::shared_ptr &HostPointer, image_format Format, image_sampler Sampler, const range &Range, @@ -1253,9 +1230,7 @@ class sampled_image CodeLoc, (void *)this->impl.get(), HostPointer.get(), Dimensions, detail::rangeToArray(Range).data(), Format, Sampler); } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template 1), typename = std::enable_if_t> sampled_image( diff --git a/sycl/include/sycl/interop_handle.hpp b/sycl/include/sycl/interop_handle.hpp index cd09f1b3aa223..55c47e1a6a36b 100644 --- a/sycl/include/sycl/interop_handle.hpp +++ b/sycl/include/sycl/interop_handle.hpp @@ -55,8 +55,8 @@ class interop_handle { /// a backend graph object is available for interop. __SYCL_EXPORT bool ext_codeplay_has_graph() const noexcept; + // access::target::constant_buffer is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH - /// Receives a SYCL accessor that has been defined as a requirement for the /// command group, and returns the underlying OpenCL memory object that is /// used by the SYCL runtime. If the accessor passed as parameter is not part diff --git a/sycl/include/sycl/multi_ptr.hpp b/sycl/include/sycl/multi_ptr.hpp index dc4623a336f50..8fbfbd81e4506 100644 --- a/sycl/include/sycl/multi_ptr.hpp +++ b/sycl/include/sycl/multi_ptr.hpp @@ -37,17 +37,18 @@ struct LegacyPointerTypes { // Specialization for constant_space to avoid creating a non-legacy multi_ptr // with the unsupported space. +// access::address_space::constant_space is deprecated. +__SYCL_SUPPRESS_DEPRECATED_PUSH template struct LegacyPointerTypes { - __SYCL_SUPPRESS_DEPRECATED_PUSH using decorated_type = typename detail::DecoratedType< ElementType, access::address_space::constant_space>::type; - __SYCL_SUPPRESS_DEPRECATED_POP using pointer_t = decorated_type *; using const_pointer_t = decorated_type const *; }; +__SYCL_SUPPRESS_DEPRECATED_POP // Helper to avoid instantiations of invalid non-legacy multi_ptr types. template @@ -61,18 +62,19 @@ struct LegacyReferenceTypes { // Specialization for constant_space to avoid creating a non-legacy multi_ptr // with the unsupported space. +// access::address_space::constant_space is deprecated. +__SYCL_SUPPRESS_DEPRECATED_PUSH template struct LegacyReferenceTypes { - __SYCL_SUPPRESS_DEPRECATED_PUSH using decorated_type = typename detail::DecoratedType< ElementType, access::address_space::constant_space>::type; - __SYCL_SUPPRESS_DEPRECATED_POP using reference_t = decorated_type &; using const_reference_t = decorated_type &; }; +__SYCL_SUPPRESS_DEPRECATED_POP } // namespace detail /// Provides constructors for address space qualified and non address space @@ -267,6 +269,7 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { return *this; } + // access::address_space::constant_space is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template < access::address_space OtherSpace, access::decorated OtherIsDecorated, @@ -278,9 +281,7 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { m_Pointer = detail::static_address_cast(Other.get_decorated()); return *this; } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH template < access::address_space OtherSpace, access::decorated OtherIsDecorated, typename = @@ -578,6 +579,7 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { static_assert(DecorateAddress != access::decorated::legacy); // constant_space is only supported in legacy multi_ptr. + // access::address_space::constant_space is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH static_assert(Space != access::address_space::constant_space, "SYCL 2020 multi_ptr does not support the deprecated " @@ -830,6 +832,8 @@ class multi_ptr { Accessor.get_pointer().get())) {} // Only if Space == local_space || generic_space + // target::local, constant_buffer, address_space::constant_space and + // accessor::get_pointer() are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template < int dimensions, access::mode Mode, access::placeholder isPlaceholder, @@ -841,7 +845,6 @@ class multi_ptr { isPlaceholder, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} - __SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == local_space || generic_space template @@ -849,7 +852,6 @@ class multi_ptr { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == constant_space - __SYCL_SUPPRESS_DEPRECATED_PUSH template < int dimensions, access::mode Mode, access::placeholder isPlaceholder, typename PropertyListT, access::address_space _Space = Space, @@ -859,7 +861,6 @@ class multi_ptr { isPlaceholder, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} - __SYCL_SUPPRESS_DEPRECATED_POP // The following constructors are necessary to create multi_ptr from @@ -888,7 +889,6 @@ class multi_ptr { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == local_space || generic_space and element type is const - __SYCL_SUPPRESS_DEPRECATED_PUSH template { target::local, isPlaceholder, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} - __SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == local_space || generic_space and element type is const template { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == constant_space and element type is const - __SYCL_SUPPRESS_DEPRECATED_PUSH template < int dimensions, access::mode Mode, access::placeholder isPlaceholder, typename PropertyListT, access::address_space _Space = Space, @@ -1135,6 +1133,8 @@ class multi_ptr { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == local_space || generic_space + // target::local, constant_buffer and address_space::constant_space are + // deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::mode Mode, @@ -1146,10 +1146,8 @@ class multi_ptr { access::placeholder::false_t, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} - __SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == local_space || generic_space - __SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::address_space _Space = Space, @@ -1158,10 +1156,8 @@ class multi_ptr { Space == access::address_space::local_space)>> multi_ptr(local_accessor Accessor) : multi_ptr(Accessor.get_pointer()) {} - __SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == constant_space - __SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::mode Mode, typename PropertyListT, access::address_space _Space = Space, @@ -1300,6 +1296,8 @@ class multi_ptr { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == local_space || generic_space + // target::local, constant_buffer and address_space::constant_space are + // deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::mode Mode, @@ -1311,10 +1309,8 @@ class multi_ptr { access::placeholder::false_t, PropertyListT> Accessor) : multi_ptr(Accessor.get_pointer()) {} - __SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == local_space || generic_space - __SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::address_space _Space = Space, @@ -1323,10 +1319,8 @@ class multi_ptr { Space == access::address_space::local_space)>> multi_ptr(local_accessor Accessor) : multi_ptr(Accessor.get_pointer()) {} - __SYCL_SUPPRESS_DEPRECATED_POP // Only if Space == constant_space - __SYCL_SUPPRESS_DEPRECATED_PUSH template < typename ElementType, int dimensions, access::mode Mode, typename PropertyListT, access::address_space _Space = Space, @@ -1386,6 +1380,8 @@ multi_ptr(accessor) -> multi_ptr; +// target::constant_buffer, local and address_space::constant_space are +// deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template @@ -1393,9 +1389,7 @@ multi_ptr(accessor) -> multi_ptr; -__SYCL_SUPPRESS_DEPRECATED_POP -__SYCL_SUPPRESS_DEPRECATED_PUSH template multi_ptr( diff --git a/sycl/include/sycl/nd_item.hpp b/sycl/include/sycl/nd_item.hpp index d742163a0ff84..bc54ee80764ae 100644 --- a/sycl/include/sycl/nd_item.hpp +++ b/sycl/include/sycl/nd_item.hpp @@ -72,6 +72,7 @@ template class nd_item { id Index = get_global_id(); range Extent = get_global_range(); + // nd_item::get_offset() is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH id Offset = get_offset(); __SYCL_SUPPRESS_DEPRECATED_POP @@ -202,6 +203,8 @@ template class nd_item { } nd_range get_nd_range() const noexcept { + // nd_item::get_offset() and the offset-taking nd_range constructor + // are deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH return nd_range(get_global_range(), get_local_range(), get_offset()); @@ -209,8 +212,9 @@ template class nd_item { } #ifndef __INTEL_PREVIEW_BREAKING_CHANGES - __SYCL_SUPPRESS_DEPRECATED_PUSH + // access::fence_space is deprecated. + __SYCL_SUPPRESS_DEPRECATED_PUSH __SYCL2020_DEPRECATED("use sycl::group_barrier() free function instead") void barrier([[maybe_unused]] access::fence_space accessSpace = access::fence_space::global_and_local) const { @@ -221,9 +225,6 @@ template class nd_item { #endif } - __SYCL_SUPPRESS_DEPRECATED_POP - - __SYCL_SUPPRESS_DEPRECATED_PUSH /// Executes a work-group mem-fence with memory ordering on the local address /// space, global address space or both based on the value of \p accessSpace. template diff --git a/sycl/include/sycl/reduction.hpp b/sycl/include/sycl/reduction.hpp index 4448d88d51b35..9cbfb5cdfc7dd 100644 --- a/sycl/include/sycl/reduction.hpp +++ b/sycl/include/sycl/reduction.hpp @@ -1353,12 +1353,12 @@ struct NDRangeReduction< }); }; - __SYCL_SUPPRESS_DEPRECATED_PUSH - auto device = getDeviceFromHandler(CGH); // Integrated/discrete GPUs have different faster path. For discrete GPUs // fast path requires USM device allocations though, so check for that as // well. + // info::device::host_unified_memory is deprecated. + __SYCL_SUPPRESS_DEPRECATED_PUSH if (device.get_info() || !device.has(aspect::usm_device_allocations)) Rest(Redu.getReadWriteAccessorToInitializedGroupsCounter(CGH)); diff --git a/sycl/include/sycl/sampler.hpp b/sycl/include/sycl/sampler.hpp index ecfd29a4455a1..10ded14e0ef16 100644 --- a/sycl/include/sycl/sampler.hpp +++ b/sycl/include/sycl/sampler.hpp @@ -142,6 +142,7 @@ struct image_sampler { } // namespace _V1 } // namespace sycl +// sycl::sampler is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template <> struct std::hash diff --git a/sycl/include/sycl/stream.hpp b/sycl/include/sycl/stream.hpp index b937e038b4bfc..a184c4613d4aa 100644 --- a/sycl/include/sycl/stream.hpp +++ b/sycl/include/sycl/stream.hpp @@ -97,8 +97,8 @@ using GlobalBufPtrType = typename detail::DecoratedType::type *; constexpr static int GlobalBufDim = 1; +// access::mode::atomic is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH - using GlobalOffsetAccessorT = accessor; @@ -648,8 +648,8 @@ writeFloatingPoint(GlobalBufAccessorT &GlobalFlushBuf, size_t FlushBufferSize, : 0); } +// sycl::atomic, returned by the atomic accessor, is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH - // Helper method to update offset in the global buffer atomically according to // the provided size of the data in the flush buffer. Return true if offset is // updated and false in case of overflow. diff --git a/sycl/include/sycl/sycl_span.hpp b/sycl/include/sycl/sycl_span.hpp index 565c2deaf5b7b..94071deb2ec3c 100644 --- a/sycl/include/sycl/sycl_span.hpp +++ b/sycl/include/sycl/sycl_span.hpp @@ -394,15 +394,14 @@ template class _SYCL_SPAN_TEMPLATE_VIS span { return rev_iterator(begin()); } + // sycl::byte is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH _SYCL_SPAN_INLINE_VISIBILITY span __as_bytes() const noexcept { return span{ reinterpret_cast(data()), size_bytes()}; } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH _SYCL_SPAN_INLINE_VISIBILITY span __as_writable_bytes() const noexcept { return span{ @@ -591,14 +590,13 @@ class _SYCL_SPAN_TEMPLATE_VIS span<_Tp, dynamic_extent> { return rev_iterator(begin()); } + // sycl::byte is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH _SYCL_SPAN_INLINE_VISIBILITY span __as_bytes() const noexcept { return {reinterpret_cast(data()), size_bytes()}; } - __SYCL_SUPPRESS_DEPRECATED_POP - __SYCL_SUPPRESS_DEPRECATED_PUSH _SYCL_SPAN_INLINE_VISIBILITY span __as_writable_bytes() const noexcept { return {reinterpret_cast(data()), size_bytes()}; diff --git a/sycl/test-e2e/Basic/accessor/Inputs/device_accessor.cpp b/sycl/test-e2e/Basic/accessor/Inputs/device_accessor.cpp index 844c2dbb8a7e3..530fbab0b1ad9 100644 --- a/sycl/test-e2e/Basic/accessor/Inputs/device_accessor.cpp +++ b/sycl/test-e2e/Basic/accessor/Inputs/device_accessor.cpp @@ -172,6 +172,7 @@ int main() { sycl::accessor acc_2(buf_data, cgh, sycl::range<1>(8), sycl::no_init); sycl::accessor acc_3(buf_data, cgh, sycl::range<1>(8), sycl::id<1>(1), sycl::no_init); + // sycl::read_constant is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH sycl::accessor acc_4(buf_data, cgh, sycl::read_constant); sycl::accessor acc_5(buf_data, cgh, sycl::range<1>(8), @@ -189,6 +190,7 @@ int main() { auto acc_2 = buf_data.get_access(cgh, sycl::range<1>(8), sycl::no_init); auto acc_3 = buf_data.get_access(cgh, sycl::range<1>(8), sycl::id<1>(1), sycl::no_init); + // sycl::read_constant is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH auto acc_4 = buf_data.get_access(cgh, sycl::read_constant); auto acc_5 = @@ -246,6 +248,7 @@ int main() { sycl::accessor acc_2(buf_data, sycl::range<1>(8), sycl::no_init); sycl::accessor acc_3(buf_data, sycl::range<1>(8), sycl::id<1>(1), sycl::no_init); + // sycl::read_constant is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH sycl::accessor acc_4(buf_data, sycl::read_constant); sycl::accessor acc_5(buf_data, sycl::range<1>(8), sycl::read_constant); @@ -262,6 +265,7 @@ int main() { auto acc_2 = buf_data.get_access(sycl::range<1>(8), sycl::no_init); auto acc_3 = buf_data.get_access(sycl::range<1>(8), sycl::id<1>(1), sycl::no_init); + // sycl::read_constant is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH auto acc_4 = buf_data.get_access(sycl::read_constant); auto acc_5 = buf_data.get_access(sycl::range<1>(8), sycl::read_constant); From 605ba05a5e6a5c3737f7bd5f5f1825f89ff55206 Mon Sep 17 00:00:00 2001 From: Jakub Stepnowski Date: Fri, 23 Oct 2026 19:49:37 +0200 Subject: [PATCH 06/11] [SYCL] Suppress internal deprecation warnings in headers Fix include formatting in reduction and span headers. --- sycl/include/sycl/reduction.hpp | 42 ++++++++++++++++----------------- sycl/include/sycl/sycl_span.hpp | 12 +++++----- 2 files changed, 27 insertions(+), 27 deletions(-) diff --git a/sycl/include/sycl/reduction.hpp b/sycl/include/sycl/reduction.hpp index dbeef3f612e3f..d1a2734461ee2 100644 --- a/sycl/include/sycl/reduction.hpp +++ b/sycl/include/sycl/reduction.hpp @@ -8,27 +8,27 @@ #pragma once -#include // for address_s... -#include // for local_acc... -#include // for aspect -#include // for IsValidAt... -#include // for atomic_ref -#include // for buffer -#include // for min -#include // for __SYCL_EX... -#include // for is_sgenfloat -#include // for createSyc... -#include // for id -#include // for strategy -#include // for __SYCL_SU... -#include // for make_tuple -#include // for device -#include // for event -#include // for make_erro... -#include // for queue_impl -#include // for plus, known_op... -#include // for workGroup... -#include // for reduce_ov... +#include // for address_s... +#include // for local_acc... +#include // for aspect +#include // for IsValidAt... +#include // for atomic_ref +#include // for buffer +#include // for min +#include // for __SYCL_SU... +#include // for __SYCL_EX... +#include // for is_sgenfloat +#include // for createSyc... +#include // for id +#include // for strategy +#include // for make_tuple +#include // for device +#include // for event +#include // for make_erro... +#include // for queue_impl +#include // for plus, known_op... +#include // for workGroup... +#include // for reduce_ov... #include #include // for handler #include // for getDeline... diff --git a/sycl/include/sycl/sycl_span.hpp b/sycl/include/sycl/sycl_span.hpp index 8f0df9f94ee76..32fffc57ef2eb 100644 --- a/sycl/include/sycl/sycl_span.hpp +++ b/sycl/include/sycl/sycl_span.hpp @@ -142,12 +142,12 @@ template */ -#include // for array -#include // for assert -#include // for size_t, nullptr_t, ptrdiff_t -#include // for SIZE_MAX -#include // for size, data, reverse_iterator -#include // for pointer_traits. +#include // for array +#include // for assert +#include // for size_t, nullptr_t, ptrdiff_t +#include // for SIZE_MAX +#include // for size, data, reverse_iterator +#include // for pointer_traits. #include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP #include // for enable_if_t, enable_if, remove_cv_t, false_type #include // for declval From d0979bdce8f78454871ac2e2e89db18bcf4e3e4f Mon Sep 17 00:00:00 2001 From: Jakub Stepnowski Date: Mon, 26 Oct 2026 16:45:49 +0100 Subject: [PATCH 07/11] [SYCL] Suppress internal deprecation warnings in headers Remove the unused detail helpers isTargetHostAccess and modeNeedsOldData from access.hpp. In fragment.hpp, replace the deprecated this_sub_group() with get_sub_group() and drop the suppression. --- sycl/include/sycl/access/access.hpp | 13 ------------- .../sycl/ext/oneapi/experimental/fragment.hpp | 6 +----- 2 files changed, 1 insertion(+), 18 deletions(-) diff --git a/sycl/include/sycl/access/access.hpp b/sycl/include/sycl/access/access.hpp index 1ed4cd5a9bad6..34c1d86a973ca 100644 --- a/sycl/include/sycl/access/access.hpp +++ b/sycl/include/sycl/access/access.hpp @@ -123,19 +123,6 @@ inline constexpr mode_target_tag_t namespace detail { -// access::target::host_buffer, host_image and access::mode::atomic are -// deprecated. -__SYCL_SUPPRESS_DEPRECATED_PUSH -constexpr bool isTargetHostAccess(access::target T) { - return T == access::target::host_buffer || T == access::target::host_image; -} - -constexpr bool modeNeedsOldData(access::mode m) { - return m == access::mode::read || m == access::mode::write || - m == access::mode::read_write || m == access::mode::atomic; -} -__SYCL_SUPPRESS_DEPRECATED_POP - constexpr bool modeWritesNewData(access::mode m) { return m != access::mode::read; } diff --git a/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp b/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp index 83c89e76bf20d..39256b85bf415 100644 --- a/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp +++ b/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp @@ -9,7 +9,6 @@ #pragma once #include -#include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP #include #include #include @@ -187,10 +186,7 @@ get_opportunistic_group() { #ifdef __SYCL_DEVICE_ONLY__ #if defined(__SPIR__) || defined(__SPIRV__) - // ext::oneapi::experimental::this_sub_group() is deprecated. - __SYCL_SUPPRESS_DEPRECATED_PUSH - sycl::sub_group sg = sycl::ext::oneapi::experimental::this_sub_group(); - __SYCL_SUPPRESS_DEPRECATED_POP + sycl::sub_group sg = sycl::ext::oneapi::this_work_item::get_sub_group(); sub_group_mask mask = sycl::ext::oneapi::group_ballot(sg, true); return fragment(mask, 0, 1); From adcc943ba2b83b90d1d96b7cad9234ffd7a63606 Mon Sep 17 00:00:00 2001 From: Jakub Stepnowski Date: Fri, 30 Oct 2026 12:03:23 +0100 Subject: [PATCH 08/11] [SYCL] Clean up include formatting and deprecation macro spacing Formats header includes and removes some unnecessary spacing around __SYCL_SUPPRESS_DEPRECATED_PUSH/POP --- sycl/include/sycl/access/access.hpp | 2 - sycl/include/sycl/accessor.hpp | 7 +-- sycl/include/sycl/accessor_image.hpp | 23 ++-------- .../sycl/detail/address_space_cast.hpp | 1 - sycl/include/sycl/detail/helpers.hpp | 4 +- sycl/include/sycl/detail/image_ocl_types.hpp | 1 - .../sycl/ext/oneapi/experimental/fragment.hpp | 2 - .../sycl/ext/oneapi/filter_selector.hpp | 2 - sycl/include/sycl/handler.hpp | 1 - sycl/include/sycl/interop_handle.hpp | 1 - sycl/include/sycl/multi_ptr.hpp | 10 ----- sycl/include/sycl/nd_item.hpp | 3 -- sycl/include/sycl/reduction.hpp | 43 +++++++++---------- sycl/include/sycl/stream.hpp | 2 - 14 files changed, 27 insertions(+), 75 deletions(-) diff --git a/sycl/include/sycl/access/access.hpp b/sycl/include/sycl/access/access.hpp index 34c1d86a973ca..35c194d65fad2 100644 --- a/sycl/include/sycl/access/access.hpp +++ b/sycl/include/sycl/access/access.hpp @@ -106,14 +106,12 @@ template struct mode_target_tag_t { inline constexpr mode_tag_t read_only{}; inline constexpr mode_tag_t read_write{}; inline constexpr mode_tag_t write_only{}; - // access::target::constant_buffer is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH __SYCL2020_DEPRECATED("use 'read_only' instead") inline constexpr mode_target_tag_t read_constant{}; __SYCL_SUPPRESS_DEPRECATED_POP - inline constexpr mode_target_tag_t read_only_host_task; inline constexpr mode_target_tag_t diff --git a/sycl/include/sycl/accessor.hpp b/sycl/include/sycl/accessor.hpp index 5b7bc3cc5755c..a9ff479a93c4a 100644 --- a/sycl/include/sycl/accessor.hpp +++ b/sycl/include/sycl/accessor.hpp @@ -294,7 +294,6 @@ class accessor_common { __SYCL_SUPPRESS_DEPRECATED_PUSH constexpr static bool IsHostBuf = AccessTarget == access::target::host_buffer; __SYCL_SUPPRESS_DEPRECATED_POP - constexpr static bool IsHostTask = AccessTarget == access::target::host_task; // SYCL2020 4.7.6.9.4.3 // IsPlaceHolder template parameter has no bearing on whether the accessor @@ -1743,7 +1742,6 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : getQualifiedPtr() + LinearIndex)); } __SYCL_SUPPRESS_DEPRECATED_POP - template 1)>> auto operator[](size_t Index) const { return AccessorSubscript(*this, Index); @@ -1797,7 +1795,6 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor : // accessor::has_property for runtime properties is only available in host // code. This restriction is not listed in the core spec and will be added in - // future versions. template typename std::enable_if_t< @@ -2369,8 +2366,6 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS accessor< accessor> { - __SYCL_SUPPRESS_DEPRECATED_POP - using local_acc = local_accessor_base; @@ -2416,7 +2411,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS accessor< accessor(const detail::AccessorImplPtr &Impl) : local_acc{Impl} {} #endif }; - +__SYCL_SUPPRESS_DEPRECATED_POP template class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(local_accessor) local_accessor : public local_accessor_base> @@ -562,7 +557,6 @@ class __image_array_slice__ { DataT read(const CoordT &Coords, const sampler &Smpl) const { return MBaseAcc.read(getAdjustedCoords(Coords), Smpl); } - __SYCL_SUPPRESS_DEPRECATED_POP template > MBaseAcc; - __SYCL_SUPPRESS_DEPRECATED_POP }; } // namespace detail - -// access::target::image and sycl::image are deprecated. -__SYCL_SUPPRESS_DEPRECATED_PUSH /// Image accessors. /// /// Available only when accessTarget == access::target::image. @@ -681,10 +670,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor< accessor() = default; #endif }; -__SYCL_SUPPRESS_DEPRECATED_POP -// access::target::host_image and sycl::image are deprecated. -__SYCL_SUPPRESS_DEPRECATED_PUSH /// Host image accessor. /// /// Available only when accessTarget == access::target::host_image. @@ -717,10 +703,7 @@ class __SYCL_EBO accessor // for MemorySemanticsMask -#include // for fence_space +#include // for MemorySemanticsMask +#include // for fence_space #include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP #include // for __SYCL_EXPORT #include // for memory_order diff --git a/sycl/include/sycl/detail/image_ocl_types.hpp b/sycl/include/sycl/detail/image_ocl_types.hpp index 35367c77585e1..7ed299ecaf2e5 100644 --- a/sycl/include/sycl/detail/image_ocl_types.hpp +++ b/sycl/include/sycl/detail/image_ocl_types.hpp @@ -564,7 +564,6 @@ struct opencl_image_type { opencl_image_type *; }; __SYCL_SUPPRESS_DEPRECATED_POP - template struct sampled_opencl_image_type { using type = void *; }; diff --git a/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp b/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp index 39256b85bf415..c5e65446e5842 100644 --- a/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp +++ b/sycl/include/sycl/ext/oneapi/experimental/fragment.hpp @@ -185,9 +185,7 @@ __SYCL_ALWAYS_INLINE inline fragment get_opportunistic_group() { #ifdef __SYCL_DEVICE_ONLY__ #if defined(__SPIR__) || defined(__SPIRV__) - sycl::sub_group sg = sycl::ext::oneapi::this_work_item::get_sub_group(); - sub_group_mask mask = sycl::ext::oneapi::group_ballot(sg, true); return fragment(mask, 0, 1); #elif defined(__NVPTX__) diff --git a/sycl/include/sycl/ext/oneapi/filter_selector.hpp b/sycl/include/sycl/ext/oneapi/filter_selector.hpp index 52913b6b70780..150e006cc1f91 100644 --- a/sycl/include/sycl/ext/oneapi/filter_selector.hpp +++ b/sycl/include/sycl/ext/oneapi/filter_selector.hpp @@ -54,9 +54,7 @@ class __SYCL_EXPORT_DEPRECATED( std::shared_ptr impl; filter_selector(sycl::detail::string_view filter); }; - __SYCL_SUPPRESS_DEPRECATED_POP - } // namespace ext::oneapi #ifdef __SYCL_INTERNAL_API diff --git a/sycl/include/sycl/handler.hpp b/sycl/include/sycl/handler.hpp index 97e317e48b103..65fde47318129 100644 --- a/sycl/include/sycl/handler.hpp +++ b/sycl/include/sycl/handler.hpp @@ -1232,7 +1232,6 @@ class __SYCL_EXPORT handler { constexpr int AccessTargetMask = 0x7ff; const access::target target = static_cast( type::sizes[NumArgs] & AccessTargetMask); - // access::target::local is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH if (target == target::local) { diff --git a/sycl/include/sycl/interop_handle.hpp b/sycl/include/sycl/interop_handle.hpp index 55c47e1a6a36b..51b871b89c9b3 100644 --- a/sycl/include/sycl/interop_handle.hpp +++ b/sycl/include/sycl/interop_handle.hpp @@ -88,7 +88,6 @@ class interop_handle { } __SYCL_SUPPRESS_DEPRECATED_POP - /// Receives a SYCL accessor that has been defined as a requirement for the /// command group, and returns the underlying OpenCL memory object that is /// used by the SYCL runtime. If the accessor passed as parameter is not part diff --git a/sycl/include/sycl/multi_ptr.hpp b/sycl/include/sycl/multi_ptr.hpp index 8fbfbd81e4506..4b7cfc9b5347e 100644 --- a/sycl/include/sycl/multi_ptr.hpp +++ b/sycl/include/sycl/multi_ptr.hpp @@ -41,10 +41,8 @@ struct LegacyPointerTypes { __SYCL_SUPPRESS_DEPRECATED_PUSH template struct LegacyPointerTypes { - using decorated_type = typename detail::DecoratedType< ElementType, access::address_space::constant_space>::type; - using pointer_t = decorated_type *; using const_pointer_t = decorated_type const *; }; @@ -67,10 +65,8 @@ __SYCL_SUPPRESS_DEPRECATED_PUSH template struct LegacyReferenceTypes { - using decorated_type = typename detail::DecoratedType< ElementType, access::address_space::constant_space>::type; - using reference_t = decorated_type &; using const_reference_t = decorated_type &; }; @@ -268,7 +264,6 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { m_Pointer = nullptr; return *this; } - // access::address_space::constant_space is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH template < @@ -281,7 +276,6 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { m_Pointer = detail::static_address_cast(Other.get_decorated()); return *this; } - template < access::address_space OtherSpace, access::decorated OtherIsDecorated, typename = @@ -577,7 +571,6 @@ class __SYCL_TYPE(multi_ptr) multi_ptr { std::add_pointer_t>); // Legacy has a different interface. static_assert(DecorateAddress != access::decorated::legacy); - // constant_space is only supported in legacy multi_ptr. // access::address_space::constant_space is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH @@ -1379,7 +1372,6 @@ template ) -> multi_ptr; - // target::constant_buffer, local and address_space::constant_space are // deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH @@ -1389,14 +1381,12 @@ multi_ptr(accessor) -> multi_ptr; - template multi_ptr( accessor) -> multi_ptr; __SYCL_SUPPRESS_DEPRECATED_POP - template multi_ptr(local_accessor) -> multi_ptr; diff --git a/sycl/include/sycl/nd_item.hpp b/sycl/include/sycl/nd_item.hpp index bc54ee80764ae..4700a1573a253 100644 --- a/sycl/include/sycl/nd_item.hpp +++ b/sycl/include/sycl/nd_item.hpp @@ -71,12 +71,10 @@ template class nd_item { size_t LinId = 0; id Index = get_global_id(); range Extent = get_global_range(); - // nd_item::get_offset() is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH id Offset = get_offset(); __SYCL_SUPPRESS_DEPRECATED_POP - if (1 == Dimensions) { LinId = Index[0] - Offset[0]; } else if (2 == Dimensions) { @@ -212,7 +210,6 @@ template class nd_item { } #ifndef __INTEL_PREVIEW_BREAKING_CHANGES - // access::fence_space is deprecated. __SYCL_SUPPRESS_DEPRECATED_PUSH __SYCL2020_DEPRECATED("use sycl::group_barrier() free function instead") diff --git a/sycl/include/sycl/reduction.hpp b/sycl/include/sycl/reduction.hpp index d1a2734461ee2..fb5d37deff28f 100644 --- a/sycl/include/sycl/reduction.hpp +++ b/sycl/include/sycl/reduction.hpp @@ -8,27 +8,27 @@ #pragma once -#include // for address_s... -#include // for local_acc... -#include // for aspect -#include // for IsValidAt... -#include // for atomic_ref -#include // for buffer -#include // for min -#include // for __SYCL_SU... -#include // for __SYCL_EX... -#include // for is_sgenfloat -#include // for createSyc... -#include // for id -#include // for strategy -#include // for make_tuple -#include // for device -#include // for event -#include // for make_erro... -#include // for queue_impl -#include // for plus, known_op... -#include // for workGroup... -#include // for reduce_ov... +#include // for address_s... +#include // for local_acc... +#include // for aspect +#include // for IsValidAt... +#include // for atomic_ref +#include // for buffer +#include // for min +#include // for __SYCL_SU... +#include // for __SYCL_EX... +#include // for is_sgenfloat +#include // for createSyc... +#include // for id +#include // for strategy +#include // for make_tuple +#include // for device +#include // for event +#include // for make_erro... +#include // for queue_impl +#include // for plus, known_op... +#include // for workGroup... +#include // for reduce_ov... #include #include // for handler #include // for getDeline... @@ -1365,7 +1365,6 @@ struct NDRangeReduction< Rest(Redu.getReadWriteAccessorToInitializedGroupsCounter(CGH)); else Rest(Redu.getGroupsCounterAccDiscrete(CGH)); - __SYCL_SUPPRESS_DEPRECATED_POP } }; diff --git a/sycl/include/sycl/stream.hpp b/sycl/include/sycl/stream.hpp index a184c4613d4aa..81cb41ffdc087 100644 --- a/sycl/include/sycl/stream.hpp +++ b/sycl/include/sycl/stream.hpp @@ -103,7 +103,6 @@ using GlobalOffsetAccessorT = accessor; __SYCL_SUPPRESS_DEPRECATED_POP - constexpr static access::address_space GlobalOffsetAS = TargetToAS::AS; using GlobalOffsetPtrType = @@ -668,7 +667,6 @@ inline bool updateOffset(GlobalOffsetAccessorT &GlobalOffset, } __SYCL_SUPPRESS_DEPRECATED_POP - inline void flushBuffer(GlobalOffsetAccessorT &GlobalOffset, GlobalBufAccessorT &GlobalBuf, GlobalBufAccessorT &GlobalFlushBuf, unsigned WIOffset) { From 320d704a18f14d08291a3562ec94b512fd414961 Mon Sep 17 00:00:00 2001 From: Jakub Stepnowski Date: Fri, 30 Oct 2026 12:10:57 +0100 Subject: [PATCH 09/11] [SYCL] Clean up include formatting --- sycl/include/sycl/detail/helpers.hpp | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/sycl/include/sycl/detail/helpers.hpp b/sycl/include/sycl/detail/helpers.hpp index 16aa6c9a940c6..b34ae5f2b4f1e 100644 --- a/sycl/include/sycl/detail/helpers.hpp +++ b/sycl/include/sycl/detail/helpers.hpp @@ -8,8 +8,8 @@ #pragma once -#include // for MemorySemanticsMask -#include // for fence_space +#include // for MemorySemanticsMask +#include // for fence_space #include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP #include // for __SYCL_EXPORT #include // for memory_order From 37f26f594b76708c4dd80188582cdd5c907cb2e9 Mon Sep 17 00:00:00 2001 From: Jakub Stepnowski Date: Fri, 30 Oct 2026 12:36:01 +0100 Subject: [PATCH 10/11] [SYCL] Clean up include formatting --- sycl/include/sycl/group.hpp | 22 +++++++++++----------- sycl/include/sycl/sampler.hpp | 8 ++++---- 2 files changed, 15 insertions(+), 15 deletions(-) diff --git a/sycl/include/sycl/group.hpp b/sycl/include/sycl/group.hpp index b28d836b96da1..9fb30d5ca0444 100644 --- a/sycl/include/sycl/group.hpp +++ b/sycl/include/sycl/group.hpp @@ -8,21 +8,21 @@ #pragma once -#include // for Scope, __ocl_event_t -#include // for decorated, mode, addr... +#include // for Scope, __ocl_event_t +#include // for decorated, mode, addr... #include #include // for convertToOpenCLGroupAsyncCopyPtr #include // for __SYCL_TYPE -#include // for __SYCL2020_DEPRECATED, __SY... +#include // for __SYCL2020_DEPRECATED, __SY... #include -#include // for Builder, getSPIRVMemo... -#include // for NDLoop -#include // for is_bool, change_base_type_t -#include // for device_event -#include // for id -#include // for memory_scope -#include // for decorated_global_ptr -#include // for range +#include // for Builder, getSPIRVMemo... +#include // for NDLoop +#include // for is_bool, change_base_type_t +#include // for device_event +#include // for id +#include // for memory_scope +#include // for decorated_global_ptr +#include // for range #ifndef __SYCL_DEVICE_ONLY__ #include diff --git a/sycl/include/sycl/sampler.hpp b/sycl/include/sycl/sampler.hpp index 10ded14e0ef16..65077f052337e 100644 --- a/sycl/include/sycl/sampler.hpp +++ b/sycl/include/sycl/sampler.hpp @@ -8,13 +8,13 @@ #pragma once -#include // for mode, placeholder, target -#include // for __SYCL_SPECIAL_CLASS, __SYCL_TYPE +#include // for mode, placeholder, target +#include // for __SYCL_SPECIAL_CLASS, __SYCL_TYPE #include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP #include // for __SYCL_EXPORT #include -#include // for getSyclObjImpl -#include // for property_list +#include // for getSyclObjImpl +#include // for property_list #include // for size_t #include // for shared_ptr, hash From 80db69f45034d97ee3f73d667a32626d88863fb7 Mon Sep 17 00:00:00 2001 From: Jakub Stepnowski Date: Fri, 30 Oct 2026 12:40:45 +0100 Subject: [PATCH 11/11] Revert "[SYCL] Clean up include formatting" This reverts commit 37f26f594b76708c4dd80188582cdd5c907cb2e9. Tried to enforce nicer formatting, did not pass CI. --- sycl/include/sycl/group.hpp | 22 +++++++++++----------- sycl/include/sycl/sampler.hpp | 8 ++++---- 2 files changed, 15 insertions(+), 15 deletions(-) diff --git a/sycl/include/sycl/group.hpp b/sycl/include/sycl/group.hpp index 9fb30d5ca0444..b28d836b96da1 100644 --- a/sycl/include/sycl/group.hpp +++ b/sycl/include/sycl/group.hpp @@ -8,21 +8,21 @@ #pragma once -#include // for Scope, __ocl_event_t -#include // for decorated, mode, addr... +#include // for Scope, __ocl_event_t +#include // for decorated, mode, addr... #include #include // for convertToOpenCLGroupAsyncCopyPtr #include // for __SYCL_TYPE -#include // for __SYCL2020_DEPRECATED, __SY... +#include // for __SYCL2020_DEPRECATED, __SY... #include -#include // for Builder, getSPIRVMemo... -#include // for NDLoop -#include // for is_bool, change_base_type_t -#include // for device_event -#include // for id -#include // for memory_scope -#include // for decorated_global_ptr -#include // for range +#include // for Builder, getSPIRVMemo... +#include // for NDLoop +#include // for is_bool, change_base_type_t +#include // for device_event +#include // for id +#include // for memory_scope +#include // for decorated_global_ptr +#include // for range #ifndef __SYCL_DEVICE_ONLY__ #include diff --git a/sycl/include/sycl/sampler.hpp b/sycl/include/sycl/sampler.hpp index 65077f052337e..10ded14e0ef16 100644 --- a/sycl/include/sycl/sampler.hpp +++ b/sycl/include/sycl/sampler.hpp @@ -8,13 +8,13 @@ #pragma once -#include // for mode, placeholder, target -#include // for __SYCL_SPECIAL_CLASS, __SYCL_TYPE +#include // for mode, placeholder, target +#include // for __SYCL_SPECIAL_CLASS, __SYCL_TYPE #include // for __SYCL_SUPPRESS_DEPRECATED_PUSH/POP #include // for __SYCL_EXPORT #include -#include // for getSyclObjImpl -#include // for property_list +#include // for getSyclObjImpl +#include // for property_list #include // for size_t #include // for shared_ptr, hash