Move sycl::event into future + remove <sycl::event> template param fr…

…om __future specializations

Signed-off-by: Sergey Kopienko <[email protected]>

include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl.h

@@ -241,7 +241,7 @@ struct __parallel_for_submitter<__internal::__optional_kernel_name<_Name...>>
                 __brick(__idx, __rngs...);
             });
         });
-        return __future<sycl::event>(__event);
+        return __future(std::move(__event));
     }
 };
 
@@ -371,7 +371,7 @@ struct __parallel_scan_submitter<_CustomName, __internal::__optional_kernel_name
             });
         });
 
-        return __future<sycl::event, __result_and_scratch_storage_t>(__final_event, std::move(__result_and_scratch));
+        return __future<sycl::event, __result_and_scratch_storage_t>(std::move(__final_event), std::move(__result_and_scratch));
     }
 };
 
@@ -640,7 +640,7 @@ struct __parallel_copy_if_static_single_group_submitter<_Size, _ElemsPerItem, _W
                     }
                 });
         });
-        return __future<sycl::event, __result_and_scratch_storage_t>(__event, std::move(__result));
+        return __future<sycl::event, __result_and_scratch_storage_t>(std::move(__event), std::move(__result));
     }
 };
 
@@ -696,7 +696,7 @@ __parallel_transform_scan_single_group(oneapi::dpl::__internal::__device_backend
                         /* _IsFullGroup= */ ::std::false_type, _Inclusive, _CustomName>>>()(
                     ::std::forward<_ExecutionPolicy>(__exec), std::forward<_InRng>(__in_rng),
                     std::forward<_OutRng>(__out_rng), __n, __init, __binary_op, __unary_op);
-            return __future<sycl::event, __result_and_scratch_storage_t>(__event, std::move(__dummy_result_and_scratch));
+            return __future<sycl::event, __result_and_scratch_storage_t>(std::move(__event), std::move(__dummy_result_and_scratch));
         };
         if (__n <= 16)
             return __single_group_scan_f(std::integral_constant<::std::uint16_t, 16>{});
@@ -730,7 +730,7 @@ __parallel_transform_scan_single_group(oneapi::dpl::__internal::__device_backend
             __parallel_transform_scan_dynamic_single_group_submitter<_Inclusive::value, _DynamicGroupScanKernel>()(
                 std::forward<_ExecutionPolicy>(__exec), std::forward<_InRng>(__in_rng),
                 std::forward<_OutRng>(__out_rng), __n, __init, __binary_op, __unary_op, __max_wg_size);
-        return __future<sycl::event, __result_and_scratch_storage_t>(__event, std::move(__dummy_result_and_scratch));
+        return __future<sycl::event, __result_and_scratch_storage_t>(std::move(__event), std::move(__dummy_result_and_scratch));
     }
 }
 
@@ -1618,7 +1618,7 @@ struct __parallel_partial_sort_submitter<__internal::__optional_kernel_name<_Glo
             });
         }
         // return future and extend lifetime of temporary buffer
-        return __future<sycl::event>(__event1);
+        return __future(std::move(__event1));
     }
 };
 

include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_fpga.h

@@ -75,7 +75,7 @@ struct __parallel_for_fpga_submitter<__internal::__optional_kernel_name<_Name...
                 }
             });
         });
-        return __future<sycl::event>(__event);
+        return __future(std::move(__event));
     }
 };
 

include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_histogram.h

@@ -516,7 +516,7 @@ __parallel_histogram_select_kernel(oneapi::dpl::__internal::__device_backend_tag
     // if bins fit into registers, use register private accumulation
     if (__num_bins <= __max_work_item_private_bins)
     {
-        return __future<sycl::event>(
+        return __future(
             __histogram_general_registers_local_reduction<__iters_per_work_item, __max_work_item_private_bins>(
                 __backend_tag, ::std::forward<_ExecutionPolicy>(__exec), __init_event, __work_group_size,
                 ::std::forward<_Range1>(__input), ::std::forward<_Range2>(__bins), __binhash_manager));
@@ -526,7 +526,7 @@ __parallel_histogram_select_kernel(oneapi::dpl::__internal::__device_backend_tag
                  __binhash_manager.get_required_SLM_elements() * sizeof(_extra_memory_type) <
              __local_mem_size)
     {
-        return __future<sycl::event>(__histogram_general_local_atomics<__iters_per_work_item>(
+        return __future(__histogram_general_local_atomics<__iters_per_work_item>(
             __backend_tag, ::std::forward<_ExecutionPolicy>(__exec), __init_event, __work_group_size,
             ::std::forward<_Range1>(__input), ::std::forward<_Range2>(__bins), __binhash_manager));
     }
@@ -537,7 +537,7 @@ __parallel_histogram_select_kernel(oneapi::dpl::__internal::__device_backend_tag
         // suggestion which but global memory limitations may increase this value to be able to fit the workgroup
         // private copies of the histogram bins in global memory.  No unrolling is taken advantage of here because it
         // is a runtime argument.
-        return __future<sycl::event>(__histogram_general_private_global_atomics(
+        return __future(__histogram_general_private_global_atomics(
             __backend_tag, ::std::forward<_ExecutionPolicy>(__exec), __init_event, __iters_per_work_item,
             __work_group_size, ::std::forward<_Range1>(__input), ::std::forward<_Range2>(__bins), __binhash_manager));
     }

include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_merge.h

@@ -173,7 +173,7 @@ struct __parallel_merge_submitter<_IdType, __internal::__optional_kernel_name<_N
                                __comp);
             });
         });
-        return __future<sycl::event>(__event);
+        return __future(std::move(__event));
     }
 };
 

include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_merge_sort.h

@@ -304,7 +304,7 @@ struct __parallel_sort_submitter<_IdType, __internal::__optional_kernel_name<_Le
             });
         }
 
-        return __future<sycl::event>(__event1);
+        return __future(std::move(__event1));
     }
 };
 

include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_radix_sort.h

@@ -866,7 +866,7 @@ __parallel_radix_sort(oneapi::dpl::__internal::__device_backend_tag, _ExecutionP
         }
     }
 
-    return __future<sycl::event>(__event);
+    return __future(std::move(__event));
 }
 
 } // namespace __par_backend_hetero

include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_reduce.h

@@ -154,7 +154,7 @@ struct __parallel_transform_reduce_small_submitter<_Tp, _Commutative, _VecSize,
                 });
         });
 
-        return __future<sycl::event, __result_and_scratch_storage_t>(__reduce_event, std::move(__scratch_container));
+        return __future<sycl::event, __result_and_scratch_storage_t>(std::move(__reduce_event), std::move(__scratch_container));
     }
 }; // struct __parallel_transform_reduce_small_submitter
 
@@ -268,7 +268,7 @@ struct __parallel_transform_reduce_work_group_kernel_submitter<_Tp, _Commutative
                 });
         });
 
-        return __future<sycl::event, __result_and_scratch_storage_t>(__reduce_event, std::move(__scratch_container));
+        return __future<sycl::event, __result_and_scratch_storage_t>(std::move(__reduce_event), std::move(__scratch_container));
     }
 }; // struct __parallel_transform_reduce_work_group_kernel_submitter
 
@@ -418,7 +418,7 @@ struct __parallel_transform_reduce_impl
             __n_groups = oneapi::dpl::__internal::__dpl_ceiling_div(__n, __size_per_work_group);
         } while (__n > 1);
 
-        return __future<sycl::event, __result_and_scratch_storage_t>(__reduce_event, std::move(__scratch_container));
+        return __future<sycl::event, __result_and_scratch_storage_t>(std::move(__reduce_event), std::move(__scratch_container));
     }
 }; // struct __parallel_transform_reduce_impl
 

include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_utils.h

@@ -715,9 +715,18 @@ class __future : private std::tuple<_Args...>
     using FutureType = __future<_Event, _Args...>;
     using _DataTuple = std::tuple<_Args...>;
 
-    __future(_Event __e) : __my_event(__e) {}
-    __future(_Event __e, const _DataTuple& __data) : std::tuple<_Args...>(__data), __my_event(__e) {}
-    __future(_Event __e, _DataTuple&& __data) : std::tuple<_Args...>(std::forward<_DataTuple>(__data)), __my_event(__e) {}
+    __future(const _Event& __e) : __my_event(__e) {}
+    __future(const _Event& __e, const _DataTuple& __data) : std::tuple<_Args...>(__data), __my_event(__e) {}
+    __future(const _Event& __e, _DataTuple&& __data)
+        : std::tuple<_Args...>(std::forward<_DataTuple>(__data)), __my_event(__e)
+    {
+    }
+    __future(_Event&& __e) : __my_event(std::move(__e)) {}
+    __future(_Event&& __e, const _DataTuple& __data) : std::tuple<_Args...>(__data), __my_event(std::move(__e)) {}
+    __future(_Event&& __e, _DataTuple&& __data)
+        : std::tuple<_Args...>(std::forward<_DataTuple>(__data)), __my_event(std::move(__e))
+    {
+    }
     __future(const FutureType&) = delete;
     __future(FutureType&&) = default;