From 2020d8e6e8a29719364c00e0ef84ad95a7641ce6 Mon Sep 17 00:00:00 2001
From: Andrei Elovikov <andrei.elovikov@intel.com>
Date: Mon, 2 Dec 2024 13:06:15 -0800
Subject: [PATCH 1/2] [SYCL] Enforce constraints from
 `sycl_ext_oneapi_reduction_properties`

---
 .../experimental/reduction_properties.hpp     | 48 +++++++++++++++----
 .../properties/properties_reduction.cpp       | 28 +++++++++++
 2 files changed, 68 insertions(+), 8 deletions(-)
 create mode 100644 sycl/test/extensions/properties/properties_reduction.cpp

diff --git a/sycl/include/sycl/ext/oneapi/experimental/reduction_properties.hpp b/sycl/include/sycl/ext/oneapi/experimental/reduction_properties.hpp
index a96378c522f82..2b84e0e70d890 100644
--- a/sycl/include/sycl/ext/oneapi/experimental/reduction_properties.hpp
+++ b/sycl/include/sycl/ext/oneapi/experimental/reduction_properties.hpp
@@ -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
@@ -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) {
   detail::CheckReductionIdentity<typename BufferT::value_type, BinaryOperation>(
@@ -95,7 +117,9 @@ 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);
@@ -103,8 +127,10 @@ auto reduction(T *var, BinaryOperation combiner, PropertyList 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);
@@ -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) {
@@ -122,7 +150,9 @@ 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);
@@ -130,8 +160,10 @@ auto reduction(T *var, const T &identity, BinaryOperation combiner,
   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);
diff --git a/sycl/test/extensions/properties/properties_reduction.cpp b/sycl/test/extensions/properties/properties_reduction.cpp
new file mode 100644
index 0000000000000..eb94010a1e523
--- /dev/null
+++ b/sycl/test/extensions/properties/properties_reduction.cpp
@@ -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));
+
+  // 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));
+}

From a71b1baaa8069fa8cb05d8eb9347dd6fc88b2acd Mon Sep 17 00:00:00 2001
From: Andrei Elovikov <andrei.elovikov@intel.com>
Date: Thu, 5 Dec 2024 12:45:30 -0800
Subject: [PATCH 2/2] Move SFINAE check to the return type

---
 .../experimental/reduction_properties.hpp     | 98 ++++++++++---------
 1 file changed, 54 insertions(+), 44 deletions(-)

diff --git a/sycl/include/sycl/ext/oneapi/experimental/reduction_properties.hpp b/sycl/include/sycl/ext/oneapi/experimental/reduction_properties.hpp
index 2b84e0e70d890..b4f2916a55bf1 100644
--- a/sycl/include/sycl/ext/oneapi/experimental/reduction_properties.hpp
+++ b/sycl/include/sycl/ext/oneapi/experimental/reduction_properties.hpp
@@ -103,72 +103,82 @@ inline constexpr bool is_valid_reduction_prop_list =
         ext::oneapi::experimental::detail::reduction_property_check_anchor,
         PropertyList>;
 
+template <typename BinaryOperation, typename PropertyList, typename... Args>
+auto convert_reduction_properties(BinaryOperation combiner,
+                                  PropertyList properties, Args &&...args) {
+  if constexpr (is_valid_reduction_prop_list<PropertyList>) {
+    auto WrappedOp = WrapOp(combiner, properties);
+    auto RuntimeProps = GetReductionPropertyList(properties);
+    return sycl::reduction(std::forward<Args>(args)..., WrappedOp,
+                           RuntimeProps);
+  } else {
+    // Invalid, will be disabled by SFINAE at the caller side. Make sure no hard
+    // error is emitted from here.
+  }
+}
 } // namespace detail
 
-template <typename BufferT, typename BinaryOperation, typename PropertyList,
-          typename = std::enable_if_t<
-              detail::is_valid_reduction_prop_list<PropertyList>>>
+template <typename BufferT, typename BinaryOperation, typename PropertyList>
 auto reduction(BufferT vars, handler &cgh, BinaryOperation combiner,
-               PropertyList properties) {
+               PropertyList properties)
+    -> std::enable_if_t<detail::is_valid_reduction_prop_list<PropertyList>,
+                        decltype(detail::convert_reduction_properties(
+                            combiner, properties, vars, cgh))> {
   detail::CheckReductionIdentity<typename BufferT::value_type, BinaryOperation>(
       properties);
-  auto WrappedOp = detail::WrapOp(combiner, properties);
-  auto RuntimeProps = detail::GetReductionPropertyList(properties);
-  return reduction(vars, cgh, WrappedOp, RuntimeProps);
+  return detail::convert_reduction_properties(combiner, properties, vars, cgh);
 }
 
-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) {
+template <typename T, typename BinaryOperation, typename PropertyList>
+auto reduction(T *var, BinaryOperation combiner, PropertyList properties)
+    -> std::enable_if_t<detail::is_valid_reduction_prop_list<PropertyList>,
+                        decltype(detail::convert_reduction_properties(
+                            combiner, properties, var))> {
   detail::CheckReductionIdentity<T, BinaryOperation>(properties);
-  auto WrappedOp = detail::WrapOp(combiner, properties);
-  auto RuntimeProps = detail::GetReductionPropertyList(properties);
-  return reduction(var, WrappedOp, RuntimeProps);
+  return detail::convert_reduction_properties(combiner, properties, var);
 }
 
-template <
-    typename T, size_t Extent, typename BinaryOperation, typename PropertyList,
-    typename =
-        std::enable_if_t<detail::is_valid_reduction_prop_list<PropertyList>>>
+template <typename T, size_t Extent, typename BinaryOperation,
+          typename PropertyList>
 auto reduction(span<T, Extent> vars, BinaryOperation combiner,
-               PropertyList properties) {
+               PropertyList properties)
+    -> std::enable_if_t<detail::is_valid_reduction_prop_list<PropertyList>,
+                        decltype(detail::convert_reduction_properties(
+                            combiner, properties, vars))> {
   detail::CheckReductionIdentity<T, BinaryOperation>(properties);
-  auto WrappedOp = detail::WrapOp(combiner, properties);
-  auto RuntimeProps = detail::GetReductionPropertyList(properties);
-  return reduction(vars, WrappedOp, RuntimeProps);
+  return detail::convert_reduction_properties(combiner, properties, vars);
 }
 
-template <typename BufferT, typename BinaryOperation, typename PropertyList,
-          typename = std::enable_if_t<
-              detail::is_valid_reduction_prop_list<PropertyList>>>
+template <typename BufferT, typename BinaryOperation, typename PropertyList>
 auto reduction(BufferT vars, handler &cgh,
                const typename BufferT::value_type &identity,
-               BinaryOperation combiner, PropertyList properties) {
-  auto WrappedOp = detail::WrapOp(combiner, properties);
-  auto RuntimeProps = detail::GetReductionPropertyList(properties);
-  return reduction(vars, cgh, identity, WrappedOp, RuntimeProps);
+               BinaryOperation combiner, PropertyList properties)
+    -> std::enable_if_t<detail::is_valid_reduction_prop_list<PropertyList>,
+                        decltype(detail::convert_reduction_properties(
+                            combiner, properties, vars, cgh, identity))> {
+  return detail::convert_reduction_properties(combiner, properties, vars, cgh,
+                                              identity);
 }
 
-template <typename T, typename BinaryOperation, typename PropertyList,
-          typename = std::enable_if_t<
-              detail::is_valid_reduction_prop_list<PropertyList>>>
+template <typename T, typename BinaryOperation, typename 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);
+               PropertyList properties)
+    -> std::enable_if_t<detail::is_valid_reduction_prop_list<PropertyList>,
+                        decltype(detail::convert_reduction_properties(
+                            combiner, properties, var, identity))> {
+  return detail::convert_reduction_properties(combiner, properties, var,
+                                              identity);
 }
 
-template <
-    typename T, size_t Extent, typename BinaryOperation, typename PropertyList,
-    typename =
-        std::enable_if_t<detail::is_valid_reduction_prop_list<PropertyList>>>
+template <typename T, size_t Extent, typename BinaryOperation,
+          typename PropertyList>
 auto reduction(span<T, Extent> vars, const T &identity,
-               BinaryOperation combiner, PropertyList properties) {
-  auto WrappedOp = detail::WrapOp(combiner, properties);
-  auto RuntimeProps = detail::GetReductionPropertyList(properties);
-  return reduction(vars, identity, WrappedOp, RuntimeProps);
+               BinaryOperation combiner, PropertyList properties)
+    -> std::enable_if_t<detail::is_valid_reduction_prop_list<PropertyList>,
+                        decltype(detail::convert_reduction_properties(
+                            combiner, properties, vars, identity))> {
+  return detail::convert_reduction_properties(combiner, properties, vars,
+                                              identity);
 }
 
 } // namespace _V1