fixes for 2026.0 compiler#3476
Conversation
|
Nice work on this so far! CEV job with this branch: http://intel-ci.intel.com/f0f24793-7dcb-f1ab-af3c-a4bf010d0e2d. FYI there are still fails in oneDAL build on windows and several validation jobs on linux side |
Vika-F
left a comment
There was a problem hiding this comment.
Please revert this change in logical_or, wait for the CI results and LGTM.
constexpr static inline auto native = [](const T& a, const T& b) {
return static_cast<T>(std::logical_or<T>{}(a, b));
};
|
Please share the latest CEV run from this branch. Also please address clang-format before merging |
| constexpr static inline sycl::logical_or<T> native{}; | ||
| #else | ||
| constexpr static inline std::logical_or<T> native{}; | ||
| }; |
There was a problem hiding this comment.
Confused how this file compiled with an extra } in here?
There was a problem hiding this comment.
I guess it was under #else branch and was never built.
There was a problem hiding this comment.
Pull request overview
This PR updates SYCL code to remove deprecated compiler features and improve compatibility with the 2026.0 compiler version.
Changes:
- Replaced deprecated
ext_intel_global_device_spacewithglobal_spacein atomic operations - Updated
logical_orbinary operations to useboolaccumulator type instead offloat/double - Removed compiler version checks (
#if __SYCL_COMPILER_VERSION >= ...) for obsolete fallback code - Updated MKL SYCL library symlinks from version 5 to version 6
Reviewed changes
Copilot reviewed 30 out of 30 changed files in this pull request and generated 2 comments.
Show a summary per file
| File | Description |
|---|---|
| dev/bazel/deps/mkl.tpl.BUILD | Updated MKL SYCL library symlinks from .so.5 to .so.6 |
| dev/bazel/deps/mkl.bzl | Updated MKL SYCL library symlinks from .so.5 to .so.6 |
| cpp/oneapi/dal/backend/primitives/reduction/functors.hpp | Fixed logical_or init_value and replaced deprecated address space |
| cpp/oneapi/dal/backend/atomic.hpp | Replaced deprecated ext_intel_global_device_space with global_space |
| cpp/oneapi/dal/backend/primitives/reduction/*.cpp | Added AccT template parameter for accumulator type in reduction operations |
| cpp/oneapi/dal/backend/primitives/selection/*.cpp | Removed obsolete compiler version checks |
| cpp/oneapi/dal/algo//backend/gpu/.cpp | Replaced deprecated address space and removed compiler version checks |
cpp/oneapi/dal/backend/primitives/reduction/test/reduction_rm_uniform_dpc.cpp
Outdated
Show resolved
Hide resolved
|
|
||
| array<float_t> groundtruth_cw() const { | ||
| auto res = array<float_t>::full(width_, binary_.init_value); | ||
| auto res = array<float_t>::full(width_, static_cast<float_t>(binary_.init_value)); |
There was a problem hiding this comment.
The explicit cast static_cast<float_t>(binary_.init_value) is needed because binary_.init_value is now of type bool for logical_or operations. Consider documenting this cast requirement to clarify why the conversion is necessary for future maintainers.
| @@ -102,7 +102,6 @@ sycl::event select_indexed_local(sycl::queue& q, | |||
| ? (bid + 1) * block | |||
| : src_count; | |||
| sycl::global_ptr<const Type> global((const Type*)(src_ptr + from)); | |||
There was a problem hiding this comment.
global is now unused after removing the #else branch. Please remove this variable (or use it in the async_work_group_copy call) to avoid unused-variable warnings that can break builds when warnings are treated as errors.
| sycl::global_ptr<const Type> global((const Type*)(src_ptr + from)); |
| sycl::local_ptr<const Float> local( | ||
| (const Float*)cache_.template get_multi_ptr<sycl::access::decorated::yes>().get_raw()); | ||
| #else | ||
| sycl::local_ptr<const Float> local((const Float*)cache_.get_pointer().get()); | ||
| #endif | ||
| Float acc = (override_init_ || (loc_idx != 0)) ? // | ||
| binary_.init_value | ||
| : output_[col_idx]; | ||
| AccT acc = (override_init_ || (loc_idx != 0)) ? // | ||
| binary_.init_value |
There was a problem hiding this comment.
After removing the compiler-version branches, both local (lines 50-51) and global (line 58) are no longer used. Please remove these variables or update the copy call to use them, to avoid unused-variable warnings.
| sycl::global_ptr<const Float> global(from); | ||
| const auto count = | ||
| sycl::min(static_cast<std::int32_t>(lm), static_cast<std::int32_t>(height_ - j)); | ||
| #if __SYCL_COMPILER_VERSION >= 20230828 | ||
| it.async_work_group_copy( | ||
| cache_.template get_multi_ptr<sycl::access::decorated::yes>(), | ||
| sycl::address_space_cast<sycl::access::address_space::global_space, | ||
| sycl::access::decorated::yes>(from), | ||
| count, | ||
| lstride_) | ||
| .wait(); |
There was a problem hiding this comment.
After removing the compiler-version branches, both local (lines 50-51) and global (line 58) are no longer used. Please remove these variables or update the copy call to use them, to avoid unused-variable warnings.
| "lib/libmkl_sycl_blas.so.6", | ||
| "lib/libmkl_sycl_lapack.so.6", | ||
| "lib/libmkl_sycl_sparse.so.6", | ||
| "lib/libmkl_sycl_dft.so.6", | ||
| "lib/libmkl_sycl_vm.so.6", | ||
| "lib/libmkl_sycl_rng.so.6", | ||
| "lib/libmkl_sycl_stats.so.6", | ||
| "lib/libmkl_sycl_data_fitting.so.6", |
There was a problem hiding this comment.
Hard-coding MKL SYCL SONAMEs to .so.6 can make Bazel builds fail on environments that still ship .so.5 (or that only provide the unversioned .so symlinks). Consider relying on the unversioned .so entries only, or making the versioned entries tolerant (e.g., via conditional selection, repository variants, or a single canonical SONAME) so the dependency is not tied to a specific MKL major at build time.
| "lib/libmkl_sycl_blas.so.6", | |
| "lib/libmkl_sycl_lapack.so.6", | |
| "lib/libmkl_sycl_sparse.so.6", | |
| "lib/libmkl_sycl_dft.so.6", | |
| "lib/libmkl_sycl_vm.so.6", | |
| "lib/libmkl_sycl_rng.so.6", | |
| "lib/libmkl_sycl_stats.so.6", | |
| "lib/libmkl_sycl_data_fitting.so.6", |
| template <typename Float, typename AccT, typename BinaryOp, typename UnaryOp> | ||
| class kernel_reduction_rm_rw_wide; | ||
|
|
||
| template <typename Float, typename BinaryOp, typename UnaryOp> | ||
| template <typename Float, typename AccT, typename BinaryOp, typename UnaryOp> |
There was a problem hiding this comment.
Introducing AccT as a required template parameter for these primitives increases the surface area for callers and can cause churn where the accumulator is simply Float. Consider giving AccT a default (e.g., typename AccT = Float) to preserve existing instantiation patterns and simplify usage, while still allowing bool accumulators for logical reductions.
| template <typename Float, typename AccT, typename BinaryOp, typename UnaryOp> | ||
| class kernel_reduction_rm_rw_blocking; | ||
|
|
||
| template <typename Float, typename BinaryOp, typename UnaryOp> | ||
| template <typename Float, typename AccT, typename BinaryOp, typename UnaryOp> |
There was a problem hiding this comment.
Introducing AccT as a required template parameter for these primitives increases the surface area for callers and can cause churn where the accumulator is simply Float. Consider giving AccT a default (e.g., typename AccT = Float) to preserve existing instantiation patterns and simplify usage, while still allowing bool accumulators for logical reductions.
| template <typename Float, typename AccT, typename BinaryOp, typename UnaryOp> | ||
| class kernel_reduction_rm_rw_narrow; | ||
|
|
||
| template <typename Float, typename BinaryOp, typename UnaryOp> | ||
| template <typename Float, typename AccT, typename BinaryOp, typename UnaryOp> |
There was a problem hiding this comment.
Introducing AccT as a required template parameter for these primitives increases the surface area for callers and can cause churn where the accumulator is simply Float. Consider giving AccT a default (e.g., typename AccT = Float) to preserve existing instantiation patterns and simplify usage, while still allowing bool accumulators for logical reductions.
| }; | ||
|
|
||
| template <typename Float, typename BinaryOp, typename UnaryOp> | ||
| template <typename Float, typename AccT, typename BinaryOp, typename UnaryOp> |
There was a problem hiding this comment.
Introducing AccT as a required template parameter for these primitives increases the surface area for callers and can cause churn where the accumulator is simply Float. Consider giving AccT a default (e.g., typename AccT = Float) to preserve existing instantiation patterns and simplify usage, while still allowing bool accumulators for logical reductions.
| struct logical_or { | ||
| using tag_t = reduce_binary_op_tag; | ||
| constexpr static inline T init_value = false; | ||
| constexpr static inline T init_value = T(false); |
There was a problem hiding this comment.
The logical_or struct is parameterized by type T, but logical OR operations should only work with boolean types. The init_value should be false (without cast), and T should be constrained to bool. The current implementation allows instantiation with non-boolean types (like float or double), which is semantically incorrect for logical operations.
PR Description
Summary
This PR updates SYCL code to remove deprecated features and improve compatibility with recent compiler versions.
Changes
Replacement of deprecated
address_spaceext_intel_global_device_spacehas been removed in recent compiler versions.Update of binary operations with
is_logicalflagis_logicalflag and updatedinit_valuetoT(false)sycl::reduce_over_group(it.get_group(), local_accum, binary.native);Removal of old compiler version checks
#if __SYCL_COMPILER_VERSION >= 20230828Notes
Checklist:
Completeness and readability
Testing
Performance