| 162 | |
| 163 | template <class ForwardIt1, class ForwardIt2> |
| 164 | sycl::event copy(sycl::queue &q, ForwardIt1 first, ForwardIt1 last, |
| 165 | ForwardIt2 d_first, const std::vector<sycl::event> &deps = {}) { |
| 166 | |
| 167 | auto size = std::distance(first, last); |
| 168 | if(size == 0) |
| 169 | return sycl::event{}; |
| 170 | |
| 171 | using value_type1 = typename std::iterator_traits<ForwardIt1>::value_type; |
| 172 | using value_type2 = typename std::iterator_traits<ForwardIt2>::value_type; |
| 173 | |
| 174 | if (std::is_trivially_copyable_v<value_type1> && |
| 175 | std::is_same_v<value_type1, value_type2> && |
| 176 | util::is_contiguous<ForwardIt1>() && util::is_contiguous<ForwardIt2>() && |
| 177 | detail::should_use_memcpy(q.get_device())) { |
| 178 | return q.memcpy(&(*d_first), &(*first), size * sizeof(value_type1), deps); |
| 179 | } else { |
| 180 | return q.parallel_for(sycl::range{size}, deps, |
| 181 | [=](sycl::id<1> id) { |
| 182 | auto input = first; |
| 183 | auto output = d_first; |
| 184 | std::advance(input, id[0]); |
| 185 | std::advance(output, id[0]); |
| 186 | *output = *input; |
| 187 | }); |
| 188 | } |
| 189 | } |
| 190 | |
| 191 | template <class ForwardIt1, class ForwardIt2, class UnaryPredicate> |
| 192 | sycl::event copy_if(sycl::queue &q, util::allocation_group &scratch_allocations, |