Skip to content

[SYCL] Enforce constraints from sycl_ext_oneapi_reduction_properties #16238

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
merged 2 commits into from
Dec 9, 2024
Merged

[SYCL] Enforce constraints from sycl_ext_oneapi_reduction_properties #16238

Show file tree
Hide file tree
Changes from 1 commit
Commits
Show all changes
2 commits
Select commit Hold shift + click to select a range
2020d8e
[SYCL] Enforce constraints from `sycl_ext_oneapi_reduction_properties`
aelovikov-intel Dec 2, 2024
a71b1ba
Move SFINAE check to the return type
aelovikov-intel Dec 5, 2024
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
48 changes: 40 additions & 8 deletions sycl/include/sycl/ext/oneapi/experimental/reduction_properties.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -32,6 +32,20 @@ struct initialize_to_identity_key
};
inline constexpr initialize_to_identity_key::value_t initialize_to_identity;

namespace detail {
struct reduction_property_check_anchor {};
} // namespace detail

template <>
struct is_property_key_of<deterministic_key,
detail::reduction_property_check_anchor>
: std::true_type {};

template <>
struct is_property_key_of<initialize_to_identity_key,
detail::reduction_property_check_anchor>
: std::true_type {};

} // namespace experimental
} // namespace oneapi
} // namespace ext
Expand Down Expand Up @@ -83,9 +97,17 @@ template <typename BinaryOperation>
struct IsDeterministicOperator<DeterministicOperatorWrapper<BinaryOperation>>
: std::true_type {};

template <typename PropertyList>
inline constexpr bool is_valid_reduction_prop_list =
ext::oneapi::experimental::detail::all_are_properties_of_v<
ext::oneapi::experimental::detail::reduction_property_check_anchor,
PropertyList>;

} // namespace detail

template <typename BufferT, typename BinaryOperation, typename PropertyList>
template <typename BufferT, typename BinaryOperation, typename PropertyList,
typename = std::enable_if_t<
detail::is_valid_reduction_prop_list<PropertyList>>>
auto reduction(BufferT vars, handler &cgh, BinaryOperation combiner,
PropertyList properties) {
Copy link
Contributor

@Pennycook Pennycook Dec 3, 2024

Choose a reason for hiding this comment

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

I'm a little worried by this pattern, because it adds a template argument that is not present in the specification. I know this works, but I don't if it's legal from a SYCL specification perspective.

I'd like to hear @gmlueck's opinion about this.

Copy link
Contributor

@gmlueck gmlueck Dec 3, 2024

Choose a reason for hiding this comment

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

In this case, it seems like you could avoid adding a new template parameter by instead adding enable_if to the handler parameter.

In general, implementing constraints seems a bit hacky in C++17 because you need to use SFINAE, and this sometimes requires adding a new template parameter or even changing a non-templated member function into a templated one. I don't know if we have a strict set of rules about whether this is completely legal. We know that changing a conversion operator from non-templated to templated changes the semantics, so we don't do this. (@aelovikov-intel has used a "mix-in" technique to avoid this for the vec class.) For the other cases, I'd advise that we avoid adding new template parameters if there is an alternative.

Copy link
Contributor Author

@aelovikov-intel aelovikov-intel Dec 3, 2024

Choose a reason for hiding this comment

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

by instead adding enable_if to the handler parameter.

I don't remember seeing such an approach anywhere in the codebase, do you have some pointers?

For the other cases, I'd advise that we avoid adding new template parameters if there is an alternative.

I thought we're adding that extra parameter almost in every corner of the project right now...

Copy link
Contributor

@gmlueck gmlueck Dec 3, 2024

Choose a reason for hiding this comment

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

Can't you do something like this:

template <typename BufferT, typename BinaryOperation, typename PropertyList>
auto reduction(BufferT vars, std::enable_if<detail::is_valid_reduction_prop_list<PropertyList>, handler>::type &cgh, BinaryOperation combiner,
               PropertyList properties)

Copy link
Contributor Author

@aelovikov-intel aelovikov-intel Dec 3, 2024

Choose a reason for hiding this comment

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

That isn't the question. What I'm saying is that I'm not aware of any place in the codebase where we use such pattern instead of extra template parameter. Why should we do anything differently here?

Copy link
Contributor

@gmlueck gmlueck Dec 3, 2024

Choose a reason for hiding this comment

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

I guess it's just general concern that adding another template parameter might cause some code to break.

@Pennycook, @rolandschulz: Didn't we see some previous example where adding a new template parameter to accessor caused something to break, even though the template parameter had a default value? Maybe I'm misremembering?

Copy link
Contributor

@Pennycook Pennycook Dec 4, 2024

Choose a reason for hiding this comment

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

That isn't the question. What I'm saying is that I'm not aware of any place in the codebase where we use such pattern instead of extra template parameter. Why should we do anything differently here?

FWIW, there are quite a few places (e.g., in the group algorithms) where we use enable_if on the return type instead of using an extra template parameter:

llvm/sycl/include/sycl/ext/oneapi/experimental/user_defined_reductions.hpp

Lines 46 to 49 in e9cbb87

// ---- reduce_over_group
template <typename GroupHelper, typename T, typename BinaryOperation>
std::enable_if_t<(is_group_helper_v<GroupHelper>), T>
reduce_over_group(GroupHelper group_helper, T x, BinaryOperation binary_op) {

Didn't we see some previous example where adding a new template parameter to accessor caused something to break, even though the template parameter had a default value? Maybe I'm misremembering?

I think you're referring to this issue: #10057. Even if the same problem can't actually be triggered for this overload, the fact we hit this unexpected issue with accessor does make me think that it would be safer to avoid adding extra templates just in case.

Copy link
Contributor

@steffenlarsen steffenlarsen Dec 4, 2024

Choose a reason for hiding this comment

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

I agree with @aelovikov-intel, this is not an uncommon pattern for SFINAE on public APIs. If we believe this to be problematic, we may have to address this in other places as well, but the solution of SFINAE'ing an argument assumes that the methods have arguments, which holds in this case, but may not hold for all.

Copy link
Contributor

@Pennycook Pennycook Dec 4, 2024

Choose a reason for hiding this comment

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

I might be wrong about this, but if we wanted to be consistent, couldn't we always implement SFINAE using the return type? Every function has a return type (but sometimes that type is void).

I appreciate it's a bit ugly to employ in a case like this where the implementation was previously relying on a deduced return type, but I think it's doable. Is there a reason that we couldn't implement this as something like:

template <typename BufferT, typename BinaryOperation, typename PropertyList>
std::enable_if_t<detail::is_valid_reduction_prop_list<PropertyList>,
                 detail::reduction_t<BufferT, BinaryOperation, PropertyList>>
reduction(BufferT vars, handler &cgh, BinaryOperation combiner, PropertyList properties) { ... }

As an added bonus, all the SFINAE would be in a place where we could legally replace it with a requires statement when we switch to C++20:

template <typename BufferT, typename BinaryOperation, typename PropertyList>
requires ValidReductionPropertyList<PropertyList>
auto reduction(BufferT vars, handler &cgh, BinaryOperation combiner, PropertyList properties) { ... }

Copy link
Contributor Author

@aelovikov-intel aelovikov-intel Dec 4, 2024

Choose a reason for hiding this comment

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

I personally prefer SFINAE on return type, with the exceptions like this one when it's deduced :) ...and ctors.

Copy link
Contributor

@steffenlarsen steffenlarsen Dec 4, 2024

Choose a reason for hiding this comment

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

I'm tempted to say I would prefer a static_assert on whether the properties are valid. That way we can control the message telling the user why they're not allowed to make the call they're trying to do.

Copy link
Contributor

@Pennycook Pennycook Dec 4, 2024

Choose a reason for hiding this comment

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

I don't that's valid in this case. The specification says that this is a Constraint, and ISO C++ says (emphasis mine):

[Note 1: Failure to meet such a condition results in the function's silent non-viability. — end note]

We could do this if we changed from a Constraint to a Mandate, but I don't know enough about why to favor one over the other.

detail::CheckReductionIdentity<typename BufferT::value_type, BinaryOperation>(
Expand All @@ -95,16 +117,20 @@ auto reduction(BufferT vars, handler &cgh, BinaryOperation combiner,
return reduction(vars, cgh, WrappedOp, RuntimeProps);
}

template <typename T, typename BinaryOperation, typename PropertyList>
template <typename T, typename BinaryOperation, typename PropertyList,
typename = std::enable_if_t<
detail::is_valid_reduction_prop_list<PropertyList>>>
auto reduction(T *var, BinaryOperation combiner, PropertyList properties) {
detail::CheckReductionIdentity<T, BinaryOperation>(properties);
auto WrappedOp = detail::WrapOp(combiner, properties);
auto RuntimeProps = detail::GetReductionPropertyList(properties);
return reduction(var, WrappedOp, RuntimeProps);
}

template <typename T, size_t Extent, typename BinaryOperation,
typename PropertyList>
template <
typename T, size_t Extent, typename BinaryOperation, typename PropertyList,
typename =
std::enable_if_t<detail::is_valid_reduction_prop_list<PropertyList>>>
auto reduction(span<T, Extent> vars, BinaryOperation combiner,
PropertyList properties) {
detail::CheckReductionIdentity<T, BinaryOperation>(properties);
Expand All @@ -113,7 +139,9 @@ auto reduction(span<T, Extent> vars, BinaryOperation combiner,
return reduction(vars, WrappedOp, RuntimeProps);
}

template <typename BufferT, typename BinaryOperation, typename PropertyList>
template <typename BufferT, typename BinaryOperation, typename PropertyList,
typename = std::enable_if_t<
detail::is_valid_reduction_prop_list<PropertyList>>>
auto reduction(BufferT vars, handler &cgh,
const typename BufferT::value_type &identity,
BinaryOperation combiner, PropertyList properties) {
Expand All @@ -122,16 +150,20 @@ auto reduction(BufferT vars, handler &cgh,
return reduction(vars, cgh, identity, WrappedOp, RuntimeProps);
}

template <typename T, typename BinaryOperation, typename PropertyList>
template <typename T, typename BinaryOperation, typename PropertyList,
typename = std::enable_if_t<
detail::is_valid_reduction_prop_list<PropertyList>>>
auto reduction(T *var, const T &identity, BinaryOperation combiner,
PropertyList properties) {
auto WrappedOp = detail::WrapOp(combiner, properties);
auto RuntimeProps = detail::GetReductionPropertyList(properties);
return reduction(var, identity, WrappedOp, RuntimeProps);
}

template <typename T, size_t Extent, typename BinaryOperation,
typename PropertyList>
template <
typename T, size_t Extent, typename BinaryOperation, typename PropertyList,
typename =
std::enable_if_t<detail::is_valid_reduction_prop_list<PropertyList>>>
auto reduction(span<T, Extent> vars, const T &identity,
BinaryOperation combiner, PropertyList properties) {
auto WrappedOp = detail::WrapOp(combiner, properties);
Expand Down
28 changes: 28 additions & 0 deletions sycl/test/extensions/properties/properties_reduction.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -0,0 +1,28 @@
// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note %s

#include <sycl/sycl.hpp>

int main() {
int *r = nullptr;
// Must not use `sycl_ext_oneapi_reduction_properties`'s overloads:
std::ignore =
sycl::reduction(r, sycl::plus<int>{},
sycl::property::reduction::initialize_to_identity{});

namespace sycl_exp = sycl::ext::oneapi::experimental;
std::ignore =
sycl::reduction(r, sycl::plus<int>{},
sycl_exp::properties(sycl_exp::initialize_to_identity));
Comment on lines +13 to +15
Copy link
Contributor

@uditagarwal97 uditagarwal97 Dec 6, 2024

Choose a reason for hiding this comment

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

Should we augment the test for deterministic property as well?

Copy link
Contributor Author

@aelovikov-intel aelovikov-intel Dec 6, 2024

Choose a reason for hiding this comment

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

Why? The change here is about "negative" behavior, this particular line acts less like a test/check and more like "context"/"back-reference". Positive behavior tests are expected to have been added in the original PR introducing these properties.

Copy link
Contributor

@uditagarwal97 uditagarwal97 Dec 9, 2024

Choose a reason for hiding this comment

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

I see. Looks good then. Also, please update the PR description to summarize the changes in this PR.

Copy link
Contributor Author

@aelovikov-intel aelovikov-intel Dec 9, 2024

Choose a reason for hiding this comment

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

I think title already says it all. What else do you want to see there?

Copy link
Contributor

@uditagarwal97 uditagarwal97 Dec 9, 2024

Choose a reason for hiding this comment

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

It is not clear from the title what "Constraints"/"Negative behavior" you are referring to.

Copy link
Contributor Author

@aelovikov-intel aelovikov-intel Dec 9, 2024

Choose a reason for hiding this comment

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

There we go. All constraints is the keyword I'm looking for.

That's probably language barrier :) In the absence of "some" I'd just assume "all" by default :)

What's the downside of quoting the constraints...

You haven't answered my question:

What would you do if different methods had slightly different constraints? Would you list all of them separately?

I can't imagine the answer for that would be copy-pasting them all into this PR. And if we agree that we wouldn't do that for such more complex case, I can't see how a simpler one deserves a longer PR description.

Copy link
Contributor

@uditagarwal97 uditagarwal97 Dec 9, 2024

Choose a reason for hiding this comment

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

What would you do if different methods had slightly different constraints? Would you list all of them separately?

Well, if the constraints were to be slightly different, we could still try to summarize them, without being overly specific.
To quote LLVM's developer policy on commit messages (https://llvm.org/docs/DeveloperPolicy.html#commit-messages):

Most importantly, the contents of the message should be carefully written to convey the rationale of the change (without delving too much in detail). It also should avoid being vague or overly specific.

Copy link
Contributor Author

@aelovikov-intel aelovikov-intel Dec 9, 2024

Choose a reason for hiding this comment

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

we could still try to summarize them

That would be "implement constraints from ", wouldn't it?

Copy link
Contributor

@uditagarwal97 uditagarwal97 Dec 9, 2024

Choose a reason for hiding this comment

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

That would be "implement constraints from ", wouldn't it?

Depends on how different the constraints are. If they are just "slightly" different, I would have come up with a better commit message than "implement constraints". If they are very different, to the extent of being unrelated, I would have tried to break PRs into smaller ones.

For the record, this is not a blocker from me since changes in SYCL RT looks good. So, if you feel strongly about not adding a commit message, please feel free to proceed with the merge.

Copy link
Contributor Author

@aelovikov-intel aelovikov-intel Dec 9, 2024

Choose a reason for hiding this comment

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

So, if you feel strongly about not adding a commit message, please feel free to proceed with the merge.

I feel strongly about [NOT] quoting the spec :)


// Not a property list:
// expected-error@+2 {{no matching function for call to 'reduction'}}
std::ignore =
sycl::reduction(r, sycl::plus<int>{}, sycl_exp::initialize_to_identity);

// Not a reduction property:
// expected-error@+2 {{no matching function for call to 'reduction'}}
std::ignore =
sycl::reduction(r, sycl::plus<int>{},
sycl_exp::properties(sycl_exp::initialize_to_identity,
sycl_exp::full_group));
}
Loading