Skip to content

[SYCL][CUDA] Revert https://github.com/intel/llvm/pull/6386 and remove c++17 usage. #6400

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.

Sign up for GitHub

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

Merged
bader merged 7 commits into from
Jul 11, 2022
Merged

[SYCL][CUDA] Revert https://github.com/intel/llvm/pull/6386 and remove c++17 usage. #6400

Show file tree
Hide file tree
Changes from 2 commits
Commits
Show all changes
7 commits
Select commit Hold shift + click to select a range
8db7dd5
Add exp builtins header to sycl.hpp and c++17 guards.
JackAKirk Jul 5, 2022
4a7b44e
remove exp builtins header include from tests.
JackAKirk Jul 5, 2022
dd5bc26
Removed `if constexpr` usage.
JackAKirk Jul 11, 2022
1a9f631
Format.
JackAKirk Jul 11, 2022
2c6814f
Format.
JackAKirk Jul 11, 2022
b417939
Remove c++17 usage from exper headers.
JackAKirk Jul 11, 2022
e924e1e
if constexpr is required for joint_matrix* functions.
JackAKirk Jul 11, 2022
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
1 change: 1 addition & 0 deletions sycl/include/CL/sycl.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -61,6 +61,7 @@
#include <sycl/ext/oneapi/backend/level_zero.hpp>
#endif
#include <sycl/ext/oneapi/device_global/properties.hpp>
#include <sycl/ext/oneapi/experimental/builtins.hpp>
#include <sycl/ext/oneapi/experimental/cuda/barrier.hpp>
#include <sycl/ext/oneapi/filter_selector.hpp>
#include <sycl/ext/oneapi/group_algorithm.hpp>
Expand Down
29 changes: 24 additions & 5 deletions sycl/include/sycl/ext/oneapi/experimental/builtins.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -143,8 +143,12 @@ sycl::marray<bfloat16, N> fabs(sycl::marray<bfloat16, N> x) {
auto partial_res = __clc_fabs(detail::to_uint32_t(x, i * 2));
std::memcpy(&res[i * 2], &partial_res, sizeof(uint32_t));
}

if constexpr (N % 2) {
#if __cplusplus >= 201703L
if constexpr (N % 2)
#else
if (N % 2)
#endif // __cplusplus >= 201703L
{
res[N - 1] = bfloat16::from_bits(__clc_fabs(x[N - 1].raw()));
}
return res;
Expand Down Expand Up @@ -179,7 +183,12 @@ sycl::marray<bfloat16, N> fmin(sycl::marray<bfloat16, N> x,
std::memcpy(&res[i * 2], &partial_res, sizeof(uint32_t));
}

if constexpr (N % 2) {
#if __cplusplus >= 201703L
if constexpr (N % 2)
#else
if (N % 2)
#endif // __cplusplus >= 201703L
{

@smaslov-intel smaslov-intel Jul 6, 2022

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

why do we need to be this explicit that this is a constexpr? Is this a performance critical place? Why not just remove constexpr?

@JackAKirk JackAKirk Jul 7, 2022

Copy link
Copy Markdown
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I think that all math functions count as performance critical places since they will often be called many many times when used. I think that we could remove constexpr here since at least at O3 there is no difference in the final asm code whether or not constexpr is explicitly used. However I was requested in the review to use c++17 where possible.

Do you want me to remove constexpr usage here?

@smaslov-intel smaslov-intel Jul 10, 2022

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I'd remove it, but I'll let you to make final call.

@JackAKirk JackAKirk Jul 11, 2022

Copy link
Copy Markdown
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

OK I've removed it. Thanks

@JackAKirk JackAKirk Jul 11, 2022

Copy link
Copy Markdown
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Sorry - I've made further commits to remove some C++17 usage consistent with #6415. if constexpr usage is necessary in joint_matrix_tensorcore.hpp.

res[N - 1] =
bfloat16::from_bits(__clc_fmin(x[N - 1].raw(), y[N - 1].raw()));
}
Expand Down Expand Up @@ -217,7 +226,12 @@ sycl::marray<bfloat16, N> fmax(sycl::marray<bfloat16, N> x,
std::memcpy(&res[i * 2], &partial_res, sizeof(uint32_t));
}

if constexpr (N % 2) {
#if __cplusplus >= 201703L
if constexpr (N % 2)
#else
if (N % 2)
#endif // __cplusplus >= 201703L
{
res[N - 1] =
bfloat16::from_bits(__clc_fmax(x[N - 1].raw(), y[N - 1].raw()));
}
Expand Down Expand Up @@ -257,7 +271,12 @@ sycl::marray<bfloat16, N> fma(sycl::marray<bfloat16, N> x,
std::memcpy(&res[i * 2], &partial_res, sizeof(uint32_t));
}

if constexpr (N % 2) {
#if __cplusplus >= 201703L
if constexpr (N % 2)
#else
if (N % 2)
#endif // __cplusplus >= 201703L
{
res[N - 1] = bfloat16::from_bits(
__clc_fma(x[N - 1].raw(), y[N - 1].raw(), z[N - 1].raw()));
}
Expand Down
1 change: 0 additions & 1 deletion sycl/test/basic_tests/built-ins.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -7,7 +7,6 @@
// Hits an assertion with AMD:
// XFAIL: hip_amd

#include <sycl/ext/oneapi/experimental/builtins.hpp>
#include <sycl/sycl.hpp>

#include <cassert>
Expand Down
1 change: 0 additions & 1 deletion sycl/test/extensions/experimental-printf.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -16,7 +16,6 @@
// CHECK: Constant [[#TYPE]] [[#CONST:]]
// CHECK: ExtInst [[#]] [[#]] [[#]] printf [[#]] [[#CONST]]

#include <sycl/ext/oneapi/experimental/builtins.hpp>
#include <sycl/sycl.hpp>

#ifdef __SYCL_DEVICE_ONLY__
Expand Down