text: /*
   *  Copyright 2008-2010 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.
   */
  
  
  /*! \file device_reference.h
   *  \brief Defines the interface to a reference to
   *         a variable which resides on a CUDA device.
   */
  
  #pragma once
  
  #include <thrust/detail/config.h>
  #include <thrust/device_ptr.h>
  #include <thrust/detail/type_traits.h>
  #include <thrust/iterator/iterator_traits.h>
  
  namespace thrust
  {
  
  /*! \addtogroup memory_management_classes Memory Management Classes
   *  \ingroup memory_management
   *  \{
   */
  
  /*! \p device_reference acts as a reference-like object to an object stored in device memory.
   *  \p device_reference is not intended to be used directly; rather, this type
   *  is the result of deferencing a \p device_ptr. Similarly, taking the address of
   *  a \p device_reference yields a \p device_ptr.
   *  
   *  \p device_reference may often be used from host code in place of operations defined on
   *  its associated \c value_type. For example, when \p device_reference refers to an
   *  arithmetic type, arithmetic operations on it are legal:
   *
   *  \code
   *  #include <thrust/device_vector.h>
   *
   *  int main(void)
   *  {
   *    thrust::device_vector<int> vec(1, 13);
   *
   *    thrust::device_reference<int> ref_to_thirteen = vec[0];
   *
   *    int x = ref_to_thirteen + 1;
   *
   *    // x is 14
   *
   *    return 0;
   *  }
   *  \endcode
   *
   *  Similarly, we can print the value of \c ref_to_thirteen in the above code by using an
   *  \c iostream:
   *
   *  \code
   *  #include <thrust/device_vector.h>
   *  #include <iostream>
   *
   *  int main(void)
   *  {
   *    thrust::device_vector<int> vec(1, 13);
   *
   *    thrust::device_reference<int> ref_to_thirteen = vec[0];
   *
   *    std::cout << ref_to_thirteen << std::endl;
   *
   *    // 13 is printed
   *
   *    return 0;
   *  }
   *  \endcode
   *
   *  Of course, we needn't explicitly create a \p device_reference in the previous
   *  example, because one is returned by \p device_vector's bracket operator. A more natural
   *  way to print the value of a \p device_vector element might be:
   *
   *  \code
   *  #include <thrust/device_vector.h>
   *  #include <iostream>
   *
   *  int main(void)
   *  {
   *    thrust::device_vector<int> vec(1, 13);
   *
   *    std::cout << vec[0] << std::endl;
   *
   *    // 13 is printed
   *
   *    return 0;
   *  }
   *  \endcode
   *
   *  These kinds of operations should be used sparingly in performance-critical code, because
   *  they imply a potentially expensive copy between host and device space.
   *
   *  Some operations which are possible with regular objects are impossible with their
   *  corresponding \p device_reference objects due to the requirements of the C++ language. For
   *  example, because the member access operator cannot be overloaded, member variables and functions
   *  of a referent object cannot be directly accessed through its \p device_reference.
   *
   *  The following code, which generates a compiler error, illustrates:
   *
   *  \code
   *  #include <thrust/device_vector.h>
   *
   *  struct foo
   *  {
   *    int x;
   *  };
   *
   *  int main(void)
   *  {
   *    thrust::device_vector<foo> foo_vec(1);
   *
   *    thrust::device_reference<foo> foo_ref = foo_vec[0];
   *
   *    foo_ref.x = 13; // ERROR: x cannot be accessed through foo_ref
   *
   *    return 0;
   *  }
   *  \endcode
   *
   *  Instead, a host space copy must be created to access \c foo's \c x member:
   *
   *  \code
   *  #include <thrust/device_vector.h>
   *
   *  struct foo
   *  {
   *    int x;
   *  };
   *
   *  int main(void)
   *  {
   *    thrust::device_vector<foo> foo_vec(1);
   *
   *    // create a local host-side foo object
   *    foo host_foo;
   *    host_foo.x = 13;
   *
   *    thrust::device_reference<foo> foo_ref = foo_vec[0];
   *
   *    foo_ref = host_foo;
   *
   *    // foo_ref's x member is 13
   *
   *    return 0;
   *  }
   *  \endcode
   *  
   *  Another common case where a \p device_reference cannot directly be used in place of
   *  its referent object occurs when passing them as parameters to functions like \c printf
   *  which have varargs parameters. Because varargs parameters must be Plain Old Data, a
   *  \p device_reference to a POD type requires a cast when passed to \c printf:
   *
   *  \code
   *  #include <stdio.h>
   *  #include <thrust/device_vector.h>
   *
   *  int main(void)
   *  {
   *    thrust::device_vector<int> vec(1,13);
   *
   *    // vec[0] must be cast to int when passing to printf
   *    printf("%d\n", (int) vec[0]);
   *
   *    return 0;
   *  }
   *  \endcode
   *
   *  \see device_ptr
   *  \see device_vector
   */
  template<typename T>
    class device_reference
  {
    public:
      typedef device_ptr<T> pointer;
      typedef typename detail::remove_const<T>::type value_type;
  
      /*! This copy constructor accepts a const reference to another
       *  \p device_reference. After this \p device_reference is constructed,
       *  it shall refer to the same object as \p ref.
       *  
       *  \param ref A \p device_reference to copy from.
       *
       *  The following code snippet demonstrates the semantics of this
       *  copy constructor.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,0);
       *  thrust::device_reference<int> ref = v[0];
       *
       *  // ref equals the object at v[0]
       *  assert(ref1 == v[0]);
       *
       *  // the address of ref equals the address of v[0]
       *  assert(&ref == &v[0]);
       *
       *  // modifying v[0] modifies ref
       *  v[0] = 13;
       *  assert(ref == 13);
       *  \endcode
       *
       *  \note This constructor is templated to initialization of const
       *  device_references to const T from device_references to T.
       */
      template<typename OtherT>
      __host__ __device__
      device_reference(const device_reference<OtherT> &ref
  
  // XXX msvc screws this up
  #ifndef _MSC_VER
                       , typename
                       detail::enable_if<
                         detail::is_convertible<
                           typename device_reference<OtherT>::pointer,
                           pointer
                         >::value
                       >::type *dummy = 0
  #endif // __MSC_VER
                      );
  
      /*! This copy constructor initializes this \p device_reference
       *  to refer to an object pointed to by the given \p device_ptr. After
       *  this \p device_reference is constructed, it shall refer to the
       *  object pointed to by \p ptr.
       *
       *  \param ptr A \p device_ptr to copy from.
       *
       *  The following code snippet demonstrates the semantic of this
       *  copy constructor.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,0);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals the object pointed to by ptr
       *  assert(ref == *ptr);
       *
       *  // the address of ref equals ptr
       *  assert(&ref == ptr);
       *
       *  // modifying *ptr modifies ref
       *  *ptr = 13;
       *  assert(ref == 13);
       *  \endcode
       */
      __host__ __device__
      explicit device_reference(const pointer &ptr);
  
      /*! Address-of operator returns a \p device_ptr pointing to the object
       *  referenced by this \p device_reference. It does not return the
       *  address of this \p device_reference.
       *
       *  \return A \p device_ptr pointing to the object this
       *  \p device_reference references.
       */
      __host__ __device__
      pointer operator&(void) const;
  
      /*! Assignment operator copies the value of the given object to the
       *  object referenced by this \p device_reference.
       *  
       *  \param v The value to copy from.
       *  \return This \p device_reference.
       */
      device_reference &operator=(const T &v);
  
      /*! Assignment operator copies the value of the object referenced by
       *  the given \p device_reference to the object referenced by this
       *  \p device_reference.
       *
       *  \param ref The \p device_reference to copy from.
       *  \return This \p device_reference.
       */
      device_reference &operator=(const device_reference &ref);
  
      /*! Assignment operator copies the value of the object referenced by
       *  the given \p device_reference to the object referenced by this
       *  \p device_reference.
       *
       *  This operator is distinguished from the previous assignment operator
       *  by being templated on the type of the other device_reference to copy from,
       *  allowing copy from a convertible type.
       *
       *  \param ref The \p device_reference to copy from.
       *  \return This \p device_reference.
       */
      template<typename OtherT>
      device_reference &operator=(const device_reference<OtherT> &ref);
  
      /*! Prefix increment operator increments the object referenced by this
       *  \p device_reference.
       *
       *  \return <tt>*this</tt>
       *  
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's prefix increment operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,0);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 0
       *  assert(ref == 0);
       *
       *  // the object pointed to by ptr equals 1
       *  assert(*ptr == 1);
       *
       *  // v[0] equals 1
       *  assert(v[0] == 1);
       *
       *  // increment ref
       *  ++ref;
       *
       *  // ref equals 1
       *  assert(ref == 1);
       *
       *  // the object pointed to by ptr equals 1
       *  assert(*ptr == 1);
       *
       *  // v[0] equals 1
       *  assert(v[0] == 1);
       *  \endcode
       *
       *  \note The increment executes as if it were executed on the host.
       *  This may change in a later version.
       */
      device_reference &operator++(void);
  
      /*! Postfix increment operator copies the object referenced by this
       *  \p device_reference, increments the object referenced by this
       *  \p device_reference, and returns the copy.
       *
       *  \return A copy of the object referenced by this \p device_reference
       *          before being incremented.
       *
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's postfix increment operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,0);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 0
       *  assert(ref == 0);
       *
       *  // the object pointed to by ptr equals 0
       *  assert(*ptr == 0);
       *
       *  // v[0] equals 0
       *  assert(v[0] == 0);
       *
       *  // increment ref
       *  int x = ref++;
       *
       *  // x equals 0
       *  assert(x == 0)
       *
       *  // ref equals 1
       *  assert(ref == 1);
       *
       *  // the object pointed to by ptr equals 1
       *  assert(*ptr == 1);
       *
       *  // v[0] equals 1
       *  assert(v[0] == 1);
       *  \endcode
       *
       *  \note The increment executes as if it were executed on the host.
       *  This may change in a later version.
       */
      value_type operator++(int);
  
      /*! Addition assignment operator add-assigns the object referenced by this
       *  \p device_reference and returns this \p device_reference.
       *
       *  \param rhs The right hand side of the add-assignment.
       *  \return <tt>*this</tt>.
       *
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's addition assignment operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,0);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 0
       *  assert(ref == 0);
       *
       *  // the object pointed to by ptr equals 0
       *  assert(*ptr == 0);
       *
       *  // v[0] equals 0
       *  assert(v[0] == 0);
       *
       *  // add-assign ref
       *  ref += 5;
       *
       *  // ref equals 5
       *  assert(ref == 5);
       *
       *  // the object pointed to by ptr equals 5
       *  assert(*ptr == 5);
       *
       *  // v[0] equals 5
       *  assert(v[0] == 5);
       *  \endcode
       *
       *  \note The add-assignment executes as as if it were executed on the host.
       *  This may change in a later version.
       */
      device_reference &operator+=(const T &rhs);
  
      /*! Prefix decrement operator decrements the object referenced by this
       *  \p device_reference.
       *
       *  \return <tt>*this</tt>
       *  
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's prefix decrement operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,0);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 0
       *  assert(ref == 0);
       *
       *  // the object pointed to by ptr equals 0
       *  assert(*ptr == 0);
       *
       *  // v[0] equals 0
       *  assert(v[0] == 0);
       *
       *  // decrement ref
       *  --ref;
       *
       *  // ref equals -1
       *  assert(ref == -1);
       *
       *  // the object pointed to by ptr equals -1
       *  assert(*ptr == -1);
       *
       *  // v[0] equals -1
       *  assert(v[0] == -1);
       *  \endcode
       *
       *  \note The decrement executes as if it were executed on the host.
       *  This may change in a later version.
       */
      device_reference &operator--(void);
  
      /*! Postfix decrement operator copies the object referenced by this
       *  \p device_reference, decrements the object referenced by this
       *  \p device_reference, and returns the copy.
       *
       *  \return A copy of the object referenced by this \p device_reference
       *          before being decremented.
       *
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's postfix decrement operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,0);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 0
       *  assert(ref == 0);
       *
       *  // the object pointed to by ptr equals 0
       *  assert(*ptr == 0);
       *
       *  // v[0] equals 0
       *  assert(v[0] == 0);
       *
       *  // decrement ref
       *  int x = ref--;
       *
       *  // x equals 0
       *  assert(x == 0)
       *
       *  // ref equals -1
       *  assert(ref == -1);
       *
       *  // the object pointed to by ptr equals -1
       *  assert(*ptr == -1);
       *
       *  // v[0] equals -1
       *  assert(v[0] == -1);
       *  \endcode
       *
       *  \note The decrement executes as if it were executed on the host.
       *  This may change in a later version.
       */
      value_type operator--(int);
  
      /*! Subtraction assignment operator subtract-assigns the object referenced by this
       *  \p device_reference and returns this \p device_reference.
       *
       *  \param rhs The right hand side of the subtraction-assignment.
       *  \return <tt>*this</tt>.
       *
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's addition assignment operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,0);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 0
       *  assert(ref == 0);
       *
       *  // the object pointed to by ptr equals 0
       *  assert(*ptr == 0);
       *
       *  // v[0] equals 0
       *  assert(v[0] == 0);
       *
       *  // subtract-assign ref
       *  ref -= 5;
       *
       *  // ref equals -5
       *  assert(ref == -5);
       *
       *  // the object pointed to by ptr equals -5
       *  assert(*ptr == -5);
       *
       *  // v[0] equals -5
       *  assert(v[0] == -5);
       *  \endcode
       *
       *  \note The subtract-assignment executes as as if it were executed on the host.
       *  This may change in a later version.
       */
      device_reference &operator-=(const T &rhs);
  
      /*! Multiplication assignment operator multiply-assigns the object referenced by this
       *  \p device_reference and returns this \p device_reference.
       *
       *  \param rhs The right hand side of the multiply-assignment.
       *  \return <tt>*this</tt>.
       *
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's multiply assignment operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,1);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 1
       *  assert(ref == 1);
       *
       *  // the object pointed to by ptr equals 1
       *  assert(*ptr == 1);
       *
       *  // v[0] equals 1
       *  assert(v[0] == 1);
       *
       *  // multiply-assign ref
       *  ref *= 5;
       *
       *  // ref equals 5
       *  assert(ref == 5);
       *
       *  // the object pointed to by ptr equals 5
       *  assert(*ptr == 5);
       *
       *  // v[0] equals 5
       *  assert(v[0] == 5);
       *  \endcode
       *
       *  \note The multiply-assignment executes as as if it were executed on the host.
       *  This may change in a later version.
       */
      device_reference &operator*=(const T &rhs);
  
      /*! Division assignment operator divide-assigns the object referenced by this
       *  \p device_reference and returns this \p device_reference.
       *
       *  \param rhs The right hand side of the divide-assignment.
       *  \return <tt>*this</tt>.
       *
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's divide assignment operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,5);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 5
       *  assert(ref == 5);
       *
       *  // the object pointed to by ptr equals 5
       *  assert(*ptr == 5);
       *
       *  // v[0] equals 5
       *  assert(v[0] == 5);
       *
       *  // divide-assign ref
       *  ref /= 5;
       *
       *  // ref equals 1
       *  assert(ref == 1);
       *
       *  // the object pointed to by ptr equals 1
       *  assert(*ptr == 1);
       *
       *  // v[0] equals 1
       *  assert(v[0] == 1);
       *  \endcode
       *
       *  \note The divide-assignment executes as as if it were executed on the host.
       *  This may change in a later version.
       */
      device_reference &operator/=(const T &rhs);
  
      /*! Modulation assignment operator modulus-assigns the object referenced by this
       *  \p device_reference and returns this \p device_reference.
       *
       *  \param rhs The right hand side of the divide-assignment.
       *  \return <tt>*this</tt>.
       *
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's divide assignment operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,5);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 5
       *  assert(ref == 5);
       *
       *  // the object pointed to by ptr equals 5
       *  assert(*ptr == 5);
       *
       *  // v[0] equals 5
       *  assert(v[0] == 5);
       *
       *  // modulus-assign ref
       *  ref %= 5;
       *
       *  // ref equals 0
       *  assert(ref == 0);
       *
       *  // the object pointed to by ptr equals 0
       *  assert(*ptr == 0);
       *
       *  // v[0] equals 0
       *  assert(v[0] == 0);
       *  \endcode
       *
       *  \note The modulus-assignment executes as as if it were executed on the host.
       *  This may change in a later version.
       */
      device_reference &operator%=(const T &rhs);
  
      /*! Bitwise left shift assignment operator left shift-assigns the object referenced by this
       *  \p device_reference and returns this \p device_reference.
       *
       *  \param rhs The right hand side of the left shift-assignment.
       *  \return <tt>*this</tt>.
       *
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's left shift assignment operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,1);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 1
       *  assert(ref == 1);
       *
       *  // the object pointed to by ptr equals 1
       *  assert(*ptr == 1);
       *
       *  // v[0] equals 1
       *  assert(v[0] == 1);
       *
       *  // left shift-assign ref
       *  ref <<= 1;
       *
       *  // ref equals 2
       *  assert(ref == 2);
       *
       *  // the object pointed to by ptr equals 2
       *  assert(*ptr == 2);
       *
       *  // v[0] equals 2
       *  assert(v[0] == 2);
       *  \endcode
       *
       *  \note The left shift-assignment executes as as if it were executed on the host.
       *  This may change in a later version.
       */
      device_reference &operator<<=(const T &rhs);
  
      /*! Bitwise right shift assignment operator right shift-assigns the object referenced by this
       *  \p device_reference and returns this \p device_reference.
       *
       *  \param rhs The right hand side of the right shift-assignment.
       *  \return <tt>*this</tt>.
       *
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's right shift assignment operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,2);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 2
       *  assert(ref == 2);
       *
       *  // the object pointed to by ptr equals 2
       *  assert(*ptr == 2);
       *
       *  // v[0] equals 2
       *  assert(v[0] == 2);
       *
       *  // right shift-assign ref
       *  ref >>= 1;
       *
       *  // ref equals 1
       *  assert(ref == 1);
       *
       *  // the object pointed to by ptr equals 1
       *  assert(*ptr == 1);
       *
       *  // v[0] equals 1
       *  assert(v[0] == 1);
       *  \endcode
       *
       *  \note The right shift-assignment executes as as if it were executed on the host.
       *  This may change in a later version.
       */
      device_reference &operator>>=(const T &rhs);
  
      /*! Bitwise AND assignment operator AND-assigns the object referenced by this
       *  \p device_reference and returns this \p device_reference.
       *
       *  \param rhs The right hand side of the AND-assignment.
       *  \return <tt>*this</tt>.
       *
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's AND assignment operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,1);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 1
       *  assert(ref == 1);
       *
       *  // the object pointed to by ptr equals 1
       *  assert(*ptr == 1);
       *
       *  // v[0] equals 1
       *  assert(v[0] == 1);
       *
       *  // right AND-assign ref
       *  ref &= 0;
       *
       *  // ref equals 0
       *  assert(ref == 0);
       *
       *  // the object pointed to by ptr equals 0
       *  assert(*ptr == 0);
       *
       *  // v[0] equals 0
       *  assert(v[0] == 0);
       *  \endcode
       *
       *  \note The AND-assignment executes as as if it were executed on the host.
       *  This may change in a later version.
       */
      device_reference &operator&=(const T &rhs);
  
      /*! Bitwise OR assignment operator OR-assigns the object referenced by this
       *  \p device_reference and returns this \p device_reference.
       *
       *  \param rhs The right hand side of the OR-assignment.
       *  \return <tt>*this</tt>.
       *
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's OR assignment operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,0);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 0
       *  assert(ref == 0);
       *
       *  // the object pointed to by ptr equals 0
       *  assert(*ptr == 0);
       *
       *  // v[0] equals 0
       *  assert(v[0] == 0);
       *
       *  // right OR-assign ref
       *  ref |= 1;
       *
       *  // ref equals 1
       *  assert(ref == 1);
       *
       *  // the object pointed to by ptr equals 1
       *  assert(*ptr == 1);
       *
       *  // v[0] equals 1
       *  assert(v[0] == 1);
       *  \endcode
       *
       *  \note The OR-assignment executes as as if it were executed on the host.
       *  This may change in a later version.
       */
      device_reference &operator|=(const T &rhs);
  
      /*! Bitwise XOR assignment operator XOR-assigns the object referenced by this
       *  \p device_reference and returns this \p device_reference.
       *
       *  \param rhs The right hand side of the XOR-assignment.
       *  \return <tt>*this</tt>.
       *
       *  The following code snippet demonstrates the semantics of
       *  \p device_reference's XOR assignment operator.
       *
       *  \code
       *  #include <thrust/device_vector.h>
       *  #include <assert.h>
       *  ...
       *  thrust::device_vector<int> v(1,1);
       *  thrust::device_ptr<int> ptr = &v[0];
       *  thrust::device_reference<int> ref(ptr);
       *
       *  // ref equals 1
       *  assert(ref == 1);
       *
       *  // the object pointed to by ptr equals 1
       *  assert(*ptr == 1);
       *
       *  // v[0] equals 1
       *  assert(v[0] == 1);
       *
       *  // right XOR-assign ref
       *  ref ^= 1;
       *
       *  // ref equals 0
       *  assert(ref == 0);
       *
       *  // the object pointed to by ptr equals 0
       *  assert(*ptr == 0);
       *
       *  // v[0] equals 0
       *  assert(v[0] == 0);
       *  \endcode
       *
       *  \note The XOR-assignment executes as as if it were executed on the host.
       *  This may change in a later version.
       */
      device_reference &operator^=(const T &rhs);
  
      /*! Conversion operator converts this \p device_reference to T
       *  by returning a copy of the object referenced by this
       *  \p device_reference.
       *
       *  \return A copy of the object referenced by this \p device_reference.
       */
      __host__ __device__
      operator value_type (void) const;
  
    private:
      // allow access to mPtr for other device_references
      template <typename OtherT> friend class device_reference;
  
      // the pointer
      const pointer mPtr;
  
      template<typename Pointer>
      inline void assign_from(Pointer src, thrust::detail::true_type spaces_are_interoperable);
  
      template<typename Pointer>
      inline void assign_from(Pointer src, thrust::detail::false_type spaces_are_not_interoperable);
  
      inline value_type convert(thrust::detail::true_type spaces_are_interoperable) const;
  
      inline value_type convert(thrust::detail::false_type spaces_are_not_interoperable) const;
  }; // end device_reference
  
  /*! \}
   */
  
  } // end thrust
  
  #include <thrust/detail/device_reference.inl>
  
  