/* * Copyright 2008-2013 NVIDIA Corporation * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. * You may obtain a copy of the License at * * http://www.apache.org/licenses/LICENSE-2.0 * * Unless required by applicable law or agreed to in writing, software * distributed under the License is distributed on an "AS IS" BASIS, * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. * See the License for the specific language governing permissions and * limitations under the License. */ #pragma once #include #if defined(_CCCL_IMPLICIT_SYSTEM_HEADER_GCC) # pragma GCC system_header #elif defined(_CCCL_IMPLICIT_SYSTEM_HEADER_CLANG) # pragma clang system_header #elif defined(_CCCL_IMPLICIT_SYSTEM_HEADER_MSVC) # pragma system_header #endif // no system header #include #include #include #include #include THRUST_NAMESPACE_BEGIN namespace detail { template OutputIterator sequential_copy(InputIterator first, InputIterator last, OutputIterator result) { for (; first != last; ++first, ++result) { *result = *first; } // end for return result; } // end sequential_copy() template BidirectionalIterator2 sequential_copy_backward(BidirectionalIterator1 first, BidirectionalIterator1 last, BidirectionalIterator2 result) { // yes, we preincrement // the ranges are open on the right, i.e. [first, last) while (first != last) { *--result = *--last; } // end while return result; } // end sequential_copy_backward() namespace dispatch { template RandomAccessIterator2 overlapped_copy( thrust::system::cpp::detail::execution_policy&, RandomAccessIterator1 first, RandomAccessIterator1 last, RandomAccessIterator2 result) { if (first < last && first <= result && result < last) { // result lies in [first, last) // it's safe to use std::copy_backward here thrust::detail::sequential_copy_backward(first, last, result + (last - first)); result += (last - first); } // end if else { // result + (last - first) lies in [first, last) // it's safe to use sequential_copy here result = thrust::detail::sequential_copy(first, last, result); } // end else return result; } // end overlapped_copy() template RandomAccessIterator2 overlapped_copy( thrust::execution_policy& exec, RandomAccessIterator1 first, RandomAccessIterator1 last, RandomAccessIterator2 result) { typedef typename thrust::iterator_value::type value_type; // make a temporary copy of [first,last), and copy into it first thrust::detail::temporary_array temp(exec, first, last); return thrust::copy(exec, temp.begin(), temp.end(), result); } // end overlapped_copy() } // namespace dispatch template RandomAccessIterator2 overlapped_copy(RandomAccessIterator1 first, RandomAccessIterator1 last, RandomAccessIterator2 result) { typedef typename thrust::iterator_system::type System1; typedef typename thrust::iterator_system::type System2; typedef typename thrust::detail::minimum_system::type System; // XXX presumes System is default constructible System system; return thrust::detail::dispatch::overlapped_copy(system, first, last, result); } // end overlapped_copy() } // namespace detail THRUST_NAMESPACE_END