TensorSyclExtractAccessor.h
00001 // This file is part of Eigen, a lightweight C++ template library
00002 // for linear algebra.
00003 //
00004 // Mehdi Goli    Codeplay Software Ltd.
00005 // Ralph Potter  Codeplay Software Ltd.
00006 // Luke Iwanski  Codeplay Software Ltd.
00007 // Contact: <eigen@codeplay.com>
00008 //
00009 // This Source Code Form is subject to the terms of the Mozilla
00010 // Public License v. 2.0. If a copy of the MPL was not distributed
00011 // with this file, You can obtain one at http://mozilla.org/MPL/2.0/.
00012 
00013 /*****************************************************************
00014  * TensorSyclExtractAccessor.h
00015  *
00016  * \brief:
00017  * ExtractAccessor takes Expression placeHolder expression and the tuple of sycl
00018  * buffers as an input. Using pre-order tree traversal, ExtractAccessor
00019  * recursively calls itself for its children in the expression tree. The
00020  * leaf node in the PlaceHolder expression is nothing but a container preserving
00021  * the order of the actual data in the tuple of sycl buffer. By invoking the
00022  * extract accessor for the PlaceHolder<N>, an accessor is created for the Nth
00023  * buffer in the tuple of buffers. This accessor is then added as an Nth
00024  * element in the tuple of accessors. In this case we preserve the order of data
00025  * in the expression tree.
00026  *
00027  * This is the specialisation of extract accessor method for different operation
00028  * type in the PlaceHolder expression.
00029  *
00030 *****************************************************************/
00031 
00032 #ifndef UNSUPPORTED_EIGEN_CXX11_SRC_TENSOR_TENSORSYCL_EXTRACT_ACCESSOR_HPP
00033 #define UNSUPPORTED_EIGEN_CXX11_SRC_TENSOR_TENSORSYCL_EXTRACT_ACCESSOR_HPP
00034 
00035 namespace Eigen {
00036 namespace TensorSycl {
00037 namespace internal {
00042 template <typename Evaluator>
00043 struct ExtractAccessor;
00044 
00045 struct AccessorConstructor{
00046   template<typename Arg> static inline auto getTuple(cl::sycl::handler& cgh, Arg eval)
00047   -> decltype(ExtractAccessor<Arg>::getTuple(cgh, eval)) {
00048   return ExtractAccessor<Arg>::getTuple(cgh, eval);
00049   }
00050 
00051   template<typename Arg1, typename Arg2> static inline auto getTuple(cl::sycl::handler& cgh, Arg1 eval1, Arg2 eval2)
00052   -> decltype(utility::tuple::append(ExtractAccessor<Arg1>::getTuple(cgh, eval1), ExtractAccessor<Arg2>::getTuple(cgh, eval2))) {
00053     return utility::tuple::append(ExtractAccessor<Arg1>::getTuple(cgh, eval1), ExtractAccessor<Arg2>::getTuple(cgh, eval2));
00054   }
00055   template<typename Arg1, typename Arg2, typename Arg3> static inline auto getTuple(cl::sycl::handler& cgh, Arg1 eval1 , Arg2 eval2 , Arg3 eval3)
00056   -> decltype(utility::tuple::append(ExtractAccessor<Arg1>::getTuple(cgh, eval1),utility::tuple::append(ExtractAccessor<Arg2>::getTuple(cgh, eval2), ExtractAccessor<Arg3>::getTuple(cgh, eval3)))) {
00057     return utility::tuple::append(ExtractAccessor<Arg1>::getTuple(cgh, eval1),utility::tuple::append(ExtractAccessor<Arg2>::getTuple(cgh, eval2), ExtractAccessor<Arg3>::getTuple(cgh, eval3)));
00058   }
00059   template< cl::sycl::access::mode AcM, typename Arg> static inline auto getAccessor(cl::sycl::handler& cgh, Arg eval)
00060   -> decltype(utility::tuple::make_tuple( eval.device().template get_sycl_accessor<AcM,
00061   typename Eigen::internal::remove_all<typename Arg::CoeffReturnType>::type>(eval.dimensions().TotalSize(), cgh,eval.data()))){
00062     return utility::tuple::make_tuple(eval.device().template get_sycl_accessor<AcM, typename Eigen::internal::remove_all<typename Arg::CoeffReturnType>::type>(eval.dimensions().TotalSize(), cgh,eval.data()));
00063   }
00064 };
00065 
00068 template <template<class, class> class UnaryCategory, typename OP, typename RHSExpr, typename Dev>
00069 struct ExtractAccessor<TensorEvaluator<const UnaryCategory<OP, RHSExpr>, Dev> > {
00070   static inline auto getTuple(cl::sycl::handler& cgh, const TensorEvaluator<const UnaryCategory<OP, RHSExpr>, Dev> eval)
00071   -> decltype(AccessorConstructor::getTuple(cgh, eval.impl())){
00072     return AccessorConstructor::getTuple(cgh, eval.impl());
00073   }
00074 };
00075 
00077 template <template<class, class> class UnaryCategory, typename OP, typename RHSExpr, typename Dev>
00078 struct ExtractAccessor<TensorEvaluator<UnaryCategory<OP, RHSExpr>, Dev> >
00079 : ExtractAccessor<TensorEvaluator<const UnaryCategory<OP, RHSExpr>, Dev> > {};
00080 
00082 template <template<class, class, class> class BinaryCategory, typename OP,  typename LHSExpr, typename RHSExpr, typename Dev>
00083 struct ExtractAccessor<TensorEvaluator<const BinaryCategory<OP, LHSExpr, RHSExpr>, Dev> > {
00084   static inline auto getTuple(cl::sycl::handler& cgh, const TensorEvaluator<const BinaryCategory<OP, LHSExpr, RHSExpr>, Dev> eval)
00085   -> decltype(AccessorConstructor::getTuple(cgh, eval.left_impl(), eval.right_impl())){
00086     return AccessorConstructor::getTuple(cgh, eval.left_impl(), eval.right_impl());
00087   }
00088 };
00090 template <template<class, class, class> class BinaryCategory, typename OP,  typename LHSExpr, typename RHSExpr, typename Dev>
00091 struct ExtractAccessor<TensorEvaluator<BinaryCategory<OP, LHSExpr, RHSExpr>, Dev> >
00092 : ExtractAccessor<TensorEvaluator<const BinaryCategory<OP, LHSExpr, RHSExpr>, Dev> >{};
00093 
00096 template <template<class, class, class, class> class TernaryCategory, typename OP, typename Arg1Expr, typename Arg2Expr, typename Arg3Expr, typename Dev>
00097 struct ExtractAccessor<TensorEvaluator<const TernaryCategory<OP, Arg1Expr, Arg2Expr, Arg3Expr>, Dev> > {
00098   static inline auto getTuple(cl::sycl::handler& cgh, const TensorEvaluator<const TernaryCategory<OP, Arg1Expr, Arg2Expr, Arg3Expr>, Dev> eval)
00099   -> decltype(AccessorConstructor::getTuple(cgh, eval.arg1Impl(), eval.arg2Impl(), eval.arg3Impl())){
00100     return AccessorConstructor::getTuple(cgh, eval.arg1Impl(), eval.arg2Impl(), eval.arg3Impl());
00101   }
00102 };
00103 
00105 template <template<class, class, class, class> class TernaryCategory, typename OP, typename Arg1Expr, typename Arg2Expr, typename Arg3Expr, typename Dev>
00106 struct ExtractAccessor<TensorEvaluator<TernaryCategory<OP, Arg1Expr, Arg2Expr, Arg3Expr>, Dev> >
00107 : ExtractAccessor<TensorEvaluator<const TernaryCategory<OP, Arg1Expr, Arg2Expr, Arg3Expr>, Dev> >{};
00108 
00111 template <typename IfExpr, typename ThenExpr, typename ElseExpr, typename Dev>
00112 struct ExtractAccessor<TensorEvaluator<const TensorSelectOp<IfExpr, ThenExpr, ElseExpr>, Dev> > {
00113   static inline auto getTuple(cl::sycl::handler& cgh, const TensorEvaluator<const TensorSelectOp<IfExpr, ThenExpr, ElseExpr>, Dev> eval)
00114   -> decltype(AccessorConstructor::getTuple(cgh, eval.cond_impl(), eval.then_impl(), eval.else_impl())){
00115     return AccessorConstructor::getTuple(cgh, eval.cond_impl(), eval.then_impl(), eval.else_impl());
00116   }
00117 };
00118 
00121 template <typename IfExpr, typename ThenExpr, typename ElseExpr, typename Dev>
00122 struct ExtractAccessor<TensorEvaluator<TensorSelectOp<IfExpr, ThenExpr, ElseExpr>, Dev> >
00123 : ExtractAccessor<TensorEvaluator<const TensorSelectOp<IfExpr, ThenExpr, ElseExpr>, Dev> >{};
00124 
00126 template <typename LHSExpr, typename RHSExpr, typename Dev>
00127 struct ExtractAccessor<TensorEvaluator<const TensorAssignOp<LHSExpr, RHSExpr>, Dev> > {
00128   static inline auto getTuple(cl::sycl::handler& cgh, const TensorEvaluator<const TensorAssignOp<LHSExpr, RHSExpr>, Dev> eval)
00129   -> decltype(AccessorConstructor::getTuple(cgh, eval.left_impl(), eval.right_impl())){
00130     return AccessorConstructor::getTuple(cgh, eval.left_impl(), eval.right_impl());
00131  }
00132 };
00133 
00135 template <typename LHSExpr, typename RHSExpr, typename Dev>
00136 struct ExtractAccessor<TensorEvaluator<TensorAssignOp<LHSExpr, RHSExpr>, Dev> >
00137 : ExtractAccessor<TensorEvaluator<const TensorAssignOp<LHSExpr, RHSExpr>, Dev> >{};
00138 
00140 #define TENSORMAPEXPR(CVQual, ACCType)\
00141 template <typename PlainObjectType, int Options_, typename Dev>\
00142 struct ExtractAccessor<TensorEvaluator<CVQual TensorMap<PlainObjectType, Options_>, Dev> > {\
00143   static inline auto getTuple(cl::sycl::handler& cgh,const TensorEvaluator<CVQual TensorMap<PlainObjectType, Options_>, Dev> eval)\
00144   -> decltype(AccessorConstructor::template getAccessor<ACCType>(cgh, eval)){\
00145     return AccessorConstructor::template getAccessor<ACCType>(cgh, eval);\
00146   }\
00147 };
00148 TENSORMAPEXPR(const, cl::sycl::access::mode::read)
00149 TENSORMAPEXPR(, cl::sycl::access::mode::read_write)
00150 #undef TENSORMAPEXPR
00151 
00153 template <typename Expr, typename Dev>
00154 struct ExtractAccessor<TensorEvaluator<const TensorForcedEvalOp<Expr>, Dev> > {
00155   static inline auto getTuple(cl::sycl::handler& cgh, const TensorEvaluator<const TensorForcedEvalOp<Expr>, Dev> eval)
00156   -> decltype(AccessorConstructor::template getAccessor<cl::sycl::access::mode::read>(cgh, eval)){
00157     return AccessorConstructor::template getAccessor<cl::sycl::access::mode::read>(cgh, eval);
00158   }
00159 };
00160 
00162 template <typename Expr, typename Dev>
00163 struct ExtractAccessor<TensorEvaluator<TensorForcedEvalOp<Expr>, Dev> >
00164 : ExtractAccessor<TensorEvaluator<const TensorForcedEvalOp<Expr>, Dev> >{};
00165 
00167 template <typename Expr, typename Dev>
00168 struct ExtractAccessor<TensorEvaluator<const TensorEvalToOp<Expr>, Dev> > {
00169   static inline auto getTuple(cl::sycl::handler& cgh,const TensorEvaluator<const TensorEvalToOp<Expr>, Dev> eval)
00170   -> decltype(utility::tuple::append(AccessorConstructor::template getAccessor<cl::sycl::access::mode::write>(cgh, eval), AccessorConstructor::getTuple(cgh, eval.impl()))){
00171     return utility::tuple::append(AccessorConstructor::template getAccessor<cl::sycl::access::mode::write>(cgh, eval), AccessorConstructor::getTuple(cgh, eval.impl()));
00172   }
00173 };
00174 
00176 template <typename Expr, typename Dev>
00177 struct ExtractAccessor<TensorEvaluator<TensorEvalToOp<Expr>, Dev> >
00178 : ExtractAccessor<TensorEvaluator<const TensorEvalToOp<Expr>, Dev> >{};
00179 
00181 template <typename OP, typename Dim, typename Expr, typename Dev>
00182 struct ExtractAccessor<TensorEvaluator<const TensorReductionOp<OP, Dim, Expr>, Dev> > {
00183   static inline auto getTuple(cl::sycl::handler& cgh, const TensorEvaluator<const TensorReductionOp<OP, Dim, Expr>, Dev> eval)
00184   -> decltype(AccessorConstructor::template getAccessor<cl::sycl::access::mode::read>(cgh, eval)){
00185     return AccessorConstructor::template getAccessor<cl::sycl::access::mode::read>(cgh, eval);
00186   }
00187 };
00188 
00190 template <typename OP, typename Dim, typename Expr, typename Dev>
00191 struct ExtractAccessor<TensorEvaluator<TensorReductionOp<OP, Dim, Expr>, Dev> >
00192 : ExtractAccessor<TensorEvaluator<const TensorReductionOp<OP, Dim, Expr>, Dev> >{};
00193 
00195 template <typename Evaluator>
00196 auto createTupleOfAccessors(cl::sycl::handler& cgh, const Evaluator& expr)
00197 -> decltype(ExtractAccessor<Evaluator>::getTuple(cgh, expr)) {
00198   return ExtractAccessor<Evaluator>::getTuple(cgh, expr);
00199 }
00200 
00201 } 
00202 } 
00203 } 
00204 #endif  // UNSUPPORTED_EIGEN_CXX11_SRC_TENSOR_TENSORSYCL_EXTRACT_ACCESSOR_HPP
 All Classes Functions Variables Typedefs Enumerator