Adding std::vector<T, usm_allocator>::iterator to is_passed_directly …

…types (#1438)


Changes USM allocator allocated std::vectors from being processed like host-side iterators to being "passed directly" (into sycl kernels) when it is possible to detect them.  The C++ standard library implementation may or may not include allocator information in the iterator type.  If it does, we can detect such iterators, and treat them as "passed directly".  Where it is not possible, they are still treated as host-side iterators and wrapped in a sycl::buffer.
Signed-off-by: Dan Hoeflinger <[email protected]>

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

@@ -149,6 +149,34 @@ struct _ModeConverter<access_mode::write>
     static constexpr access_mode __value = access_mode::discard_write;
 };
 
+template <typename Iter, typename ValueType = typename std::iterator_traits<Iter>::value_type>
+using __default_alloc_vec_iter = typename std::vector<ValueType>::iterator;
+
+template <typename Iter, typename ValueType = typename std::iterator_traits<Iter>::value_type>
+using __usm_shared_alloc_vec_iter =
+    typename std::vector<ValueType, typename sycl::usm_allocator<ValueType, sycl::usm::alloc::shared>>::iterator;
+
+template <typename Iter, typename ValueType = typename std::iterator_traits<Iter>::value_type>
+using __usm_host_alloc_vec_iter =
+    typename std::vector<ValueType, typename sycl::usm_allocator<ValueType, sycl::usm::alloc::host>>::iterator;
+
+// Evaluates to true if the provided type is an iterator with a value_type and if the implementation of a
+// std::vector<value_type, Alloc>::iterator can be distinguished between three different allocators, the
+// default, usm_shared, and usm_host. If all are distinct, it is very unlikely any non-usm based allocator
+// could be confused with a usm allocator.
+template <typename Iter, typename Void = void>
+struct __vector_iter_distinguishes_by_allocator : std::false_type
+{
+};
+template <typename Iter>
+struct __vector_iter_distinguishes_by_allocator<
+    Iter, std::enable_if_t<!std::is_same_v<__default_alloc_vec_iter<Iter>, __usm_shared_alloc_vec_iter<Iter>> &&
+                           !std::is_same_v<__default_alloc_vec_iter<Iter>, __usm_host_alloc_vec_iter<Iter>> &&
+                           !std::is_same_v<__usm_host_alloc_vec_iter<Iter>, __usm_shared_alloc_vec_iter<Iter>>>>
+    : std::true_type
+{
+};
+
 } // namespace __internal
 
 template <typename T, typename Allocator>

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

@@ -22,6 +22,7 @@
 #include "../../utils_ranges.h"
 #include "../../iterator_impl.h"
 #include "../../glue_numeric_defs.h"
+#include "sycl_iterator.h"
 #include "sycl_defs.h"
 
 namespace oneapi
@@ -206,6 +207,16 @@ struct is_passed_directly<Iter, ::std::enable_if_t<Iter::is_passed_directly::val
 {
 };
 
+//support std::vector::iterator with usm host / shared allocator as passed directly
+template <typename Iter>
+struct is_passed_directly<
+    Iter, std::enable_if_t<(std::is_same_v<Iter, oneapi::dpl::__internal::__usm_shared_alloc_vec_iter<Iter>> ||
+                            std::is_same_v<Iter, oneapi::dpl::__internal::__usm_host_alloc_vec_iter<Iter>>) &&
+                            oneapi::dpl::__internal::__vector_iter_distinguishes_by_allocator<Iter>::value>> :
+                                std::true_type
+{
+};
+
 template <typename Ip>
 struct is_passed_directly<oneapi::dpl::counting_iterator<Ip>> : ::std::true_type
 {