-
Notifications
You must be signed in to change notification settings - Fork 787
[ESIMD] Introduce atomic_update<native::lsc::fadd>(...) and similar ops. #6629
New issue
Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.
By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.
Already on GitHub? Sign in to your account
Conversation
ec6ce71
to
c6e1ba2
Compare
Signed-off-by: Konstantin S Bobrovsky <[email protected]>
c6e1ba2
to
92c00d5
Compare
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Looks good.
Most of my comments are minor and could be fixed in a separate change,
but there are 2 comments related to naming of the atomic ops that are created in non-experimental namespace and may need sooner resolution to not pollute non-experimental namespace with something to be changed later/soon.
sycl/include/sycl/ext/intel/experimental/esimd/detail/math_intrin.hpp
Outdated
Show resolved
Hide resolved
… detail/memory_intrin.hpp
@@ -357,14 +356,14 @@ std::enable_if_t<is_clang_vector_type_v<To> && is_clang_vector_type_v<From>, To> | |||
template <typename U> constexpr bool is_type() { return false; } |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
BTW, we have type checker with exactly same definition, but a better name:
https://github.com/intel/llvm/blob/sycl/sycl/include/sycl/ext/intel/experimental/esimd/detail/util.hpp#L32
sycl::ext::intel::esimd::native::lsc::atomic_op
esimd::atomic_update
is invoked withesimd::native::lsc::atomic_op
(rather thanesimd::atomic_op
), then implementation is redirected to Xe-specific LSC-based implementation (won't run until Gen12)esimd::atomic_op
cause warning and are replaced with corresponding LSC operations.Complementary E2E test: intel/llvm-test-suite#1171
Signed-off-by: Konstantin S Bobrovsky [email protected]