1 /* 2 * Copyright 2008-2013 NVIDIA Corporation 3 * 4 * Licensed under the Apache License, Version 2.0 (the "License"); 5 * you may not use this file except in compliance with the License. 6 * You may obtain a copy of the License at 7 * 8 * http://www.apache.org/licenses/LICENSE-2.0 9 * 10 * Unless required by applicable law or agreed to in writing, software 11 * distributed under the License is distributed on an "AS IS" BASIS, 12 * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. 13 * See the License for the specific language governing permissions and 14 * limitations under the License. 15 */ 16 17 18 /*! \file device_ptr.h 19 * \brief A pointer to a variable which resides in the "device" system's memory space 20 */ 21 22 #pragma once 23 24 #include <thrust/detail/config.h> 25 #include <thrust/memory.h> 26 27 namespace thrust 28 { 29 30 /*! \addtogroup memory_management Memory Management 31 * \addtogroup memory_management_classes Memory Management Classes 32 * \ingroup memory_management 33 * \{ 34 */ 35 36 // forward declarations 37 template<typename T> class device_reference; 38 39 /*! \p device_ptr stores a pointer to an object allocated in device memory. This type 40 * provides type safety when dispatching standard algorithms on ranges resident in 41 * device memory. 42 * 43 * \p device_ptr has pointer semantics: it may be dereferenced safely from the host and 44 * may be manipulated with pointer arithmetic. 45 * 46 * \p device_ptr can be created with the functions device_malloc, device_new, or 47 * device_pointer_cast, or by explicitly calling its constructor with a raw pointer. 48 * 49 * The raw pointer encapsulated by a \p device_ptr may be obtained by either its <tt>get</tt> 50 * method or the \p raw_pointer_cast free function. 51 * 52 * \note \p device_ptr is not a smart pointer; it is the programmer's responsibility to 53 * deallocate memory pointed to by \p device_ptr. 54 * 55 * \see device_malloc 56 * \see device_new 57 * \see device_pointer_cast 58 * \see raw_pointer_cast 59 */ 60 template<typename T> 61 class device_ptr 62 : public thrust::pointer< 63 T, 64 thrust::device_system_tag, 65 thrust::device_reference<T>, 66 thrust::device_ptr<T> 67 > 68 { 69 private: 70 typedef thrust::pointer< 71 T, 72 thrust::device_system_tag, 73 thrust::device_reference<T>, 74 thrust::device_ptr<T> 75 > super_t; 76 77 public: 78 /*! \p device_ptr's null constructor initializes its raw pointer to \c 0. 79 */ 80 __host__ __device__ device_ptr()81 device_ptr() : super_t() {} 82 83 #if THRUST_CPP_DIALECT >= 2011 84 // NOTE: This is needed so that Thrust smart pointers can be used in 85 // `std::unique_ptr`. 86 __host__ __device__ device_ptr(decltype (nullptr))87 device_ptr(decltype(nullptr)) : super_t(nullptr) {} 88 #endif 89 90 /*! \p device_ptr's copy constructor is templated to allow copying to a 91 * <tt>device_ptr<const T></tt> from a <tt>T *</tt>. 92 * 93 * \param ptr A raw pointer to copy from, presumed to point to a location in 94 * device memory. 95 */ 96 template<typename OtherT> 97 __host__ __device__ device_ptr(OtherT * ptr)98 explicit device_ptr(OtherT *ptr) : super_t(ptr) {} 99 100 /*! \p device_ptr's copy constructor allows copying from another device_ptr with related type. 101 * \param other The \p device_ptr to copy from. 102 */ 103 template<typename OtherT> 104 __host__ __device__ device_ptr(const device_ptr<OtherT> & other)105 device_ptr(const device_ptr<OtherT> &other) : super_t(other) {} 106 107 /*! \p device_ptr's assignment operator allows assigning from another \p device_ptr with related type. 108 * \param other The other \p device_ptr to copy from. 109 * \return <tt>*this</tt> 110 */ 111 template<typename OtherT> 112 __host__ __device__ 113 device_ptr &operator=(const device_ptr<OtherT> &other) 114 { 115 super_t::operator=(other); 116 return *this; 117 } 118 119 #if THRUST_CPP_DIALECT >= 2011 120 // NOTE: This is needed so that Thrust smart pointers can be used in 121 // `std::unique_ptr`. 122 __host__ __device__ decltype(nullptr)123 device_ptr& operator=(decltype(nullptr)) 124 { 125 super_t::operator=(nullptr); 126 return *this; 127 } 128 #endif 129 130 // declare these members for the purpose of Doxygenating them 131 // they actually exist in a derived-from class 132 #if 0 133 /*! This method returns this \p device_ptr's raw pointer. 134 * \return This \p device_ptr's raw pointer. 135 */ 136 __host__ __device__ 137 T *get(void) const; 138 #endif // end doxygen-only members 139 }; // end device_ptr 140 141 // declare these methods for the purpose of Doxygenating them 142 // they actually are defined for a derived-from class 143 #if 0 144 /*! Writes to an output stream the value of a \p device_ptr's raw pointer. 145 * 146 * \param os The output stream. 147 * \param p The \p device_ptr to output. 148 * \return os. 149 */ 150 template<typename T, typename charT, typename traits> 151 std::basic_ostream<charT, traits> & 152 operator<<(std::basic_ostream<charT, traits> &os, const device_ptr<T> &p); 153 #endif 154 155 /*! \} 156 */ 157 158 159 /*! 160 * \addtogroup memory_management_functions Memory Management Functions 161 * \ingroup memory_management 162 * \{ 163 */ 164 165 /*! \p device_pointer_cast creates a device_ptr from a raw pointer which is presumed to point 166 * to a location in device memory. 167 * 168 * \param ptr A raw pointer, presumed to point to a location in device memory. 169 * \return A device_ptr wrapping ptr. 170 */ 171 template<typename T> 172 __host__ __device__ 173 inline device_ptr<T> device_pointer_cast(T *ptr); 174 175 /*! This version of \p device_pointer_cast creates a copy of a device_ptr from another device_ptr. 176 * This version is included for symmetry with \p raw_pointer_cast. 177 * 178 * \param ptr A device_ptr. 179 * \return A copy of \p ptr. 180 */ 181 template<typename T> 182 __host__ __device__ 183 inline device_ptr<T> device_pointer_cast(const device_ptr<T> &ptr); 184 185 /*! \} 186 */ 187 188 } // end thrust 189 190 #include <thrust/detail/device_ptr.inl> 191 #include <thrust/detail/raw_pointer_cast.h> 192 193