Add tests of device_ptr and vector.
Without scatter.h cannot build, the current version that will be the next step, to implement
MODIFY
test/CMakeLists.txt
test/CMakeLists.txt
@@ -60,3 +60,4 @@ add_hipthrust_test("thrust.hip.copy" test_copy.cpp)
add_hipthrust_test("thrust.hip.for_each" test_for_each.cpp)
add_hipthrust_test("thrust.hip.memory" test_memory.cpp)
add_hipthrust_test("thrust.hip.transform" test_transform.cpp)
+add_hipthrust_test("thrust.hip.device_ptr" test_device_ptr.cpp)

ADD
test/test_device_ptr.cpp
+// MIT License
+//
+// Copyright (c) 2018 Advanced Micro Devices, Inc. All rights reserved.
+//
+// Permission is hereby granted, free of charge, to any person obtaining a copy
+// of this software and associated documentation files (the "Software"), to deal
+// in the Software without restriction, including without limitation the rights
+// to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+// copies of the Software, and to permit persons to whom the Software is
+// furnished to do so, subject to the following conditions:
+//
+// The above copyright notice and this permission notice shall be included in all
+// copies or substantial portions of the Software.
+//
+// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+// IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+// FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+// AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+// LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+// OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+// SOFTWARE.
+
+#include <iostream>
+#include <type_traits>
+#include <cstdlib>
+#include <vector>
+
+// Google Test
+#include <gtest/gtest.h>
+
+// Thrust
+#include <thrust/memory.h>
+#include <thrust/transform.h>
+// STREAMHPC TODO replace <thrust/detail/seq.h> with <thrust/execution_policy.h>
+#include <thrust/detail/seq.h>
+#include <thrust/device_ptr.h>
+#include <thrust/device_vector.h>
+
+// HIP API
+#if THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_HCC
+#include <hip/hip_runtime_api.h>
+#include <hip/hip_runtime.h>
+
+#define HIP_CHECK(condition) ASSERT_EQ(condition, hipSuccess)
+#endif // THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_HCC
+
+template<
+ class InputType
+>
+struct Params
+{
+ using input_type = InputType;
+};
+
+template<class Params>
+class DevicePtrTests : public ::testing::Test
+{
+public:
+ using input_type = typename Params::input_type;
+};
+
+typedef ::testing::Types<
+ Params<int>
+> DevicePtrTestsParams;
+
+TYPED_TEST_CASE(DevicePtrTests, DevicePtrTestsParams);
+
+#if THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_HCC
+
+TYPED_TEST(DevicePtrTests, MakeDevicePointer)
+{
+ using T = typename TestFixture::input_type;
+
+ T *raw_ptr = 0;
+ thrust::device_ptr<T> p0 = thrust::device_pointer_cast(raw_ptr);
+
+ ASSERT_EQ(thrust::raw_pointer_cast(p0), raw_ptr);
+ thrust::device_ptr<T> p1 = thrust::device_pointer_cast(p0);
+ ASSERT_EQ(p0, p1);
+}
+
+TEST(DevicePtrTests,TestDevicePointerManipulation)
+{
+ thrust::device_vector<int> data(5);
+
+ thrust::device_ptr<int> begin(&data[0]);
+ thrust::device_ptr<int> end(&data[0] + 5);
+
+ ASSERT_EQ(end - begin, 5);
+
+ begin++;
+ begin--;
+
+ ASSERT_EQ(end - begin, 5);
+
+ begin += 1;
+ begin -= 1;
+
+ ASSERT_EQ(end - begin, 5);
+
+ begin = begin + (int) 1;
+ begin = begin - (int) 1;
+
+ ASSERT_EQ(end - begin, 5);
+
+ begin = begin + (unsigned int) 1;
+ begin = begin - (unsigned int) 1;
+
+ ASSERT_EQ(end - begin, 5);
+
+ begin = begin + (size_t) 1;
+ begin = begin - (size_t) 1;
+
+ ASSERT_EQ(end - begin, 5);
+
+ begin = begin + (ptrdiff_t) 1;
+ begin = begin - (ptrdiff_t) 1;
+
+ ASSERT_EQ(end - begin, 5);
+
+ begin = begin + (thrust::device_ptr<int>::difference_type) 1;
+ begin = begin - (thrust::device_ptr<int>::difference_type) 1;
+
+ ASSERT_EQ(end - begin, 5);
+}
+
+#endif // THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_HCC

ADD
thrust/system/hip/detail/vector.inl
+/*
+ * Copyright 2008-2013 NVIDIA Corporation
+ *
+ * Licensed under the Apache License, Version 2.0 (the "License");
+ * you may not use this file except in ccudaliance 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 <thrust/detail/config.h>
+#include <thrust/system/hip/vector.h>
+
+namespace thrust
+{
+namespace hip_rocprim
+{
+
+template<typename T, typename Allocator>
+ vector<T,Allocator>
+ ::vector()
+ : super_t()
+{}
+
+template<typename T, typename Allocator>
+ vector<T,Allocator>
+ ::vector(size_type n)
+ : super_t(n)
+{}
+
+template<typename T, typename Allocator>
+ vector<T,Allocator>
+ ::vector(size_type n, const value_type &value)
+ : super_t(n,value)
+{}
+
+template<typename T, typename Allocator>
+ vector<T,Allocator>
+ ::vector(const vector &x)
+ : super_t(x)
+{}
+
+template<typename T, typename Allocator>
+ template<typename OtherT, typename OtherAllocator>
+ vector<T,Allocator>
+ ::vector(const thrust::detail::vector_base<OtherT,OtherAllocator> &x)
+ : super_t(x)
+{}
+
+template<typename T, typename Allocator>
+ template<typename OtherT, typename OtherAllocator>
+ vector<T,Allocator>
+ ::vector(const std::vector<OtherT,OtherAllocator> &x)
+ : super_t(x)
+{}
+
+template<typename T, typename Allocator>
+ template<typename InputIterator>
+ vector<T,Allocator>
+ ::vector(InputIterator first, InputIterator last)
+ : super_t(first,last)
+{}
+
+template<typename T, typename Allocator>
+ template<typename OtherT, typename OtherAllocator>
+ vector<T,Allocator> &
+ vector<T,Allocator>
+ ::operator=(const std::vector<OtherT,OtherAllocator> &x)
+{
+ super_t::operator=(x);
+ return *this;
+}
+
+template<typename T, typename Allocator>
+ template<typename OtherT, typename OtherAllocator>
+ vector<T,Allocator> &
+ vector<T,Allocator>
+ ::operator=(const thrust::detail::vector_base<OtherT,OtherAllocator> &x)
+{
+ super_t::operator=(x);
+ return *this;
+}
+
+} // end hip_rocprim
+} // end thrust

MODIFY
thrust/system/hip/vector.h
thrust/system/hip/vector.h
* Copyright 2008-2013 NVIDIA Corporation
*
* Licensed under the Apache License, Version 2.0 (the "License");
- * you may not use this file except in ccudaliance with 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
* limitations under the License.
*/
-/*! \file thrust/system/cuda_bulk/vector.h
+/*! \file thrust/system/hip/vector.h
* \brief A dynamically-sizable array of elements which reside in memory available to
- * Thrust's CUDA system.
+ * Thrust's hip system.
*/
#pragma once
@@ -38,19 +38,19 @@ namespace hip_rocprim
// XXX upon c++11
// template<typename T, typename Allocator = allocator<T> > using vector = thrust::detail::vector_base<T,Allocator>;
-/*! \p cuda_bulk::vector is a container that supports random access to elements,
+/*! \p hip_rocprim::vector is a container that supports random access to elements,
* constant time removal of elements at the end, and linear time insertion
* and removal of elements at the beginning or in the middle. The number of
- * elements in a \p cuda_bulk::vector may vary dynamically; memory management is
- * automatic. The elements contained in a \p cuda_bulk::vector reside in memory
- * available to the \p cuda_bulk system.
+ * elements in a \p hip_rocprim::vector may vary dynamically; memory management is
+ * automatic. The elements contained in a \p hip_rocprim::vector reside in memory
+ * available to the \p hip_rocprim system.
*
- * \tparam T The element type of the \p cuda_bulk::vector.
- * \tparam Allocator The allocator type of the \p cuda_bulk::vector. Defaults to \p cuda_bulk::allocator.
+ * \tparam T The element type of the \p hip_rocprim::vector.
+ * \tparam Allocator The allocator type of the \p hip_rocprim::vector. Defaults to \p hip_rocprim::allocator.
*
* \see http://www.sgi.com/tech/stl/Vector.html
* \see host_vector For the documentation of the complete interface which is
- * shared by \p cuda_bulk::vector
+ * shared by \p hip_rocprim::vector
* \see device_vector
*/
template<typename T, typename Allocator = allocator<T> >
@@ -73,23 +73,23 @@ template<typename T, typename Allocator = allocator<T> >
/*! \endcond
*/
- /*! This constructor creates an empty \p cuda_bulk::vector.
+ /*! This constructor creates an empty \p hip_rocprim::vector.
*/
vector();
- /*! This constructor creates a \p cuda_bulk::vector with \p n default-constructed elements.
- * \param n The size of the \p cuda_bulk::vector to create.
+ /*! This constructor creates a \p hip_rocprim::vector with \p n default-constructed elements.
+ * \param n The size of the \p hip_rocprim::vector to create.
*/
explicit vector(size_type n);
- /*! This constructor creates a \p cuda_bulk::vector with \p n copies of \p value.
- * \param n The size of the \p cuda_bulk::vector to create.
+ /*! This constructor creates a \p hip_rocprim::vector with \p n copies of \p value.
+ * \param n The size of the \p hip_rocprim::vector to create.
* \param value An element to copy.
*/
explicit vector(size_type n, const value_type &value);
- /*! Copy constructor copies from another \p cuda_bulk::vector.
- * \param x The other \p cuda_bulk::vector to copy.
+ /*! Copy constructor copies from another \p hip_rocprim::vector.
+ * \param x The other \p hip_rocprim::vector to copy.
*/
vector(const vector &x);
@@ -105,7 +105,7 @@ template<typename T, typename Allocator = allocator<T> >
template<typename OtherT, typename OtherAllocator>
vector(const std::vector<OtherT,OtherAllocator> &x);
- /*! This constructor creates a \p cuda_bulk::vector by copying from a range.
+ /*! This constructor creates a \p hip_rocprim::vector by copying from a range.
* \param first The beginning of the range.
* \param last The end of the range.
*/
@@ -131,21 +131,14 @@ template<typename T, typename Allocator = allocator<T> >
} // end hip_rocprim
-// alias system::cuda_bulk names at top-level
+// alias system::hip_rocprim names at top-level
namespace hip_rocprim
{
-using thrust::cuda_cub::vector;
+using thrust::hip_rocprim::vector;
} // end hip_rocprim
-namespace system {
-namespace hip {
-using thrust::hip_rocprim::vector;
-}
-}
-
} // end thrust
-#include <thrust/system/cuda/detail/vector.inl>
-
+#include <thrust/system/hip/detail/vector.inl>

Add tests of device_ptr and vector.

Without scatter.h cannot build, the current version that will be the next step, to implement