Added sycl function objects

Author: maximdimakov

sycl/include/CL/sycl/ONEAPI/functional.hpp

@@ -11,9 +11,6 @@
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
-
-template <typename T = void> using plus = std::plus<T>;
-
 namespace ONEAPI {
 
 template <typename T = void> struct minimum {

sycl/include/CL/sycl/stl.hpp

@@ -72,5 +72,48 @@ unique_ptr_class<T> make_unique_ptr(ArgsT &&... Args) {
   return unique_ptr_class<T>(new T(std::forward<ArgsT>(Args)...));
 }
 
+template <typename T = void> using plus = std::plus<T>;
+template <typename T = void> using multiplies = std::multiplies<T>;
+template <typename T = void> using bit_or = std::bit_or<T>;
+template <typename T = void> using bit_xor = std::bit_xor<T>;
+template <typename T = void> using bit_and = std::bit_and<T>;
+template <typename T = void> using logical_and = std::logical_and<T>;
+template <typename T = void> using logical_or = std::logical_or<T>;
+
+template <typename T = void> struct minimum {
+  T operator()(const T &lhs, const T &rhs) const {
+    return std::less<T>()(lhs, rhs) ? lhs : rhs;
+  }
+};
+
+template <> struct minimum<void> {
+  struct is_transparent {};
+  template <typename T, typename U>
+  auto operator()(T &&lhs, U &&rhs) const ->
+      typename std::common_type<T &&, U &&>::type {
+    return std::less<>()(std::forward<const T>(lhs), std::forward<const U>(rhs))
+               ? std::forward<T>(lhs)
+               : std::forward<U>(rhs);
+  }
+};
+
+template <typename T = void> struct maximum {
+  T operator()(const T &lhs, const T &rhs) const {
+    return std::greater<T>()(lhs, rhs) ? lhs : rhs;
+  }
+};
+
+template <> struct maximum<void> {
+  struct is_transparent {};
+  template <typename T, typename U>
+  auto operator()(T &&lhs, U &&rhs) const ->
+      typename std::common_type<T &&, U &&>::type {
+    return std::greater<>()(std::forward<const T>(lhs),
+                            std::forward<const U>(rhs))
+               ? std::forward<T>(lhs)
+               : std::forward<U>(rhs);
+  }
+};
+
 } // namespace sycl
 } // __SYCL_INLINE_NAMESPACE(cl)

sycl/test/extensions/group-algorithm.cpp

@@ -58,7 +58,7 @@ void test(queue q, InputContainer input, OutputContainer output,
             out[3] = exclusive_scan(g, in[lid], binary_op);
             out[4] = broadcast(g, in[lid]);
             out[5] = any_of(g, in.get_pointer(), in.get_pointer() + N, pred);
-            out[6] = all_of(g, pred(in[lid]));
+            out[6] = all_of_group(g, pred(in[lid]));
             if (leader(g)) {
               out[7]++;
             }
@@ -76,13 +76,13 @@ int main() {
   std::iota(input.begin(), input.end(), 0);
   std::fill(output.begin(), output.end(), 0);
 
-  test<class KernelNamePlusV>(q, input, output, sycl::plus<>(), 0, GeZero());
-  test<class KernelNameMinimumV>(q, input, output, minimum<>(),
+  test<class KernelNamePlusV>(q, input, output, ONEAPI::plus<>(), 0, GeZero());
+  test<class KernelNameMinimumV>(q, input, output, ONEAPI::minimum<>(),
                                  std::numeric_limits<int>::max(), IsEven());
 
 #ifdef SPIRV_1_3
   test<class KernelName_WonwuUVPUPOTKRKIBtT>(q, input, output,
-                                             multiplies<int>(), 1, LtZero());
+                                             ONEAPI::multiplies<int>(), 1, LtZero());
 #endif // SPIRV_1_3
 
   std::cout << "Test passed." << std::endl;