libutils/MDView.cuh

   1 /*
   2         Copyright (C) 2012  fmaj7b5.info
   3
   4         This program is free software: you can redistribute it and/or modify
   5         it under the terms of the GNU General Public License as published by
   6         the Free Software Foundation, either version 2 of the License, or
   7         (at your option) any later version.
   8
   9         This program is distributed in the hope that it will be useful,
  10         but WITHOUT ANY WARRANTY; without even the implied warranty of
  11         MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
  12         GNU General Public License for more details.
  13
  14         You should have received a copy of the GNU General Public License
  15         along with this program.  If not, see <http://www.gnu.org/licenses/>.
  16 */
  17
  18 #ifndef FM7b5_MDVIEW_CUH
  19 #define FM7b5_MDVIEW_CUH
  20
  21 #include <cstdint>
  22 #include <type_traits>
  23
  24 namespace FM7b5
  25 {
  26         namespace detail
  27         {
  28                 //! copy const qualifier of From to To
  29                 template <typename From, typename To> struct copy_const { typedef To type; };
  30                 template <typename From, typename To> struct copy_const<const From, To> { typedef typename std::add_const<To>::type type; };
  31
  32                 //! array subscript access for multi dimentional data
  33                 template <typename T, size_t D>
  34                 class MDView
  35                 {
  36                 public:
  37                         typedef T value_type;
  38                         typedef typename copy_const<T, uint_fast8_t>::type* byte_pointer_type;
  39                         typedef typename copy_const<T, void>::type* void_pointer_type;
  40
  41                 protected:
  42                         //! temporal class for resolving the operator[]
  43                         template <typename T, size_t D>
  44                         class ref
  45                         {
  46                                 byte_pointer_type const m_p;
  47                                 const size_t* const m_extent;
  48
  49                         public:
  50                                 typedef ref<T, D-1> result_type;
  51
  52                                 __device__
  53                                 ref(void_pointer_type p, const size_t* extent)
  54                                         : m_p(static_cast<byte_pointer_type>(p)), m_extent(extent)
  55                                 {}
  56
  57                                 __device__
  58                                 result_type operator[](size_t i) const
  59                                 {
  60                                         return result_type(m_p + *m_extent * i, m_extent + 1);
  61                                 }
  62                         };
  63
  64                         template <typename T>
  65                         class ref<T, 1>
  66                         {
  67                                 byte_pointer_type const m_p;
  68                                 const size_t* const m_extent;
  69
  70                         public:
  71                                 typedef typename std::add_reference<T>::type result_type;
  72
  73                                 __device__
  74                                 ref(void_pointer_type p, const size_t* extent)
  75                                         : m_p(static_cast<byte_pointer_type>(p)), m_extent(extent)
  76                                 {}
  77
  78                                 __device__
  79                                 result_type operator[](size_t i) const
  80                                 {
  81                                         return *reinterpret_cast<T*>(m_p + *m_extent * i);
  82                                 }
  83                         };
  84
  85                         byte_pointer_type m_data;
  86                         size_t m_extent[D];
  87
  88                 public:
  89                         MDView() {}
  90
  91                         MDView(void_pointer_type p, const size_t extent[D])
  92                                 : m_data(static_cast<byte_pointer_type>(p))
  93                         {
  94                                 for (size_t i = 0; i < D; ++i) {
  95                                         m_extent[i] = extent[i];
  96                                 }
  97                         }
  98
  99                         __device__
 100                         typename ref<T, D>::result_type operator[](size_t i) const
 101                         {
 102                                 ref<T, D> r(m_data, m_extent);
 103                                 return r[i];
 104                         }
 105                 };
 106         }
 107
 108         using detail::MDView;
 109 }
 110
 111 #endif /* FM7b5_MDVIEW_CUH */