10 #ifndef EIGEN_CXX11_TENSOR_TENSOR_EVALUATOR_H 11 #define EIGEN_CXX11_TENSOR_TENSOR_EVALUATOR_H 27 template<
typename Derived,
typename Device>
30 typedef typename Derived::Index Index;
31 typedef typename Derived::Scalar Scalar;
32 typedef typename Derived::Packet Packet;
33 typedef typename Derived::Scalar CoeffReturnType;
34 typedef typename Derived::Packet PacketReturnType;
35 typedef typename Derived::Dimensions Dimensions;
38 static const int NumCoords = internal::traits<Derived>::NumDimensions > 0 ?
39 internal::traits<Derived>::NumDimensions : 0;
42 IsAligned = Derived::IsAligned,
43 PacketAccess = Derived::PacketAccess,
44 Layout = Derived::Layout,
45 CoordAccess = NumCoords > 0,
48 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
TensorEvaluator(
const Derived& m,
const Device& device)
49 : m_data(const_cast<Scalar*>(m.data())), m_dims(m.dimensions()), m_device(device)
52 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
const Dimensions& dimensions()
const {
return m_dims; }
54 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
bool evalSubExprsIfNeeded(CoeffReturnType* dest) {
56 m_device.memcpy((
void*)dest, m_data,
sizeof(Scalar) * m_dims.TotalSize());
62 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
void cleanup() { }
64 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE CoeffReturnType coeff(Index index)
const {
69 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE Scalar& coeffRef(Index index) {
74 template<
int LoadMode> EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
75 PacketReturnType packet(Index index)
const 77 return internal::ploadt<Packet, LoadMode>(m_data + index);
80 template <
int StoreMode> EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
81 void writePacket(Index index,
const Packet& x)
83 return internal::pstoret<Scalar, Packet, StoreMode>(m_data + index, x);
86 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE CoeffReturnType coeff(
const array<DenseIndex, NumCoords>& coords)
const {
88 if (static_cast<int>(Layout) == static_cast<int>(ColMajor)) {
89 return m_data[m_dims.IndexOfColMajor(coords)];
91 return m_data[m_dims.IndexOfRowMajor(coords)];
95 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE Scalar& coeffRef(
const array<DenseIndex, NumCoords>& coords) {
97 if (static_cast<int>(Layout) == static_cast<int>(ColMajor)) {
98 return m_data[m_dims.IndexOfColMajor(coords)];
100 return m_data[m_dims.IndexOfRowMajor(coords)];
104 EIGEN_DEVICE_FUNC Scalar* data()
const {
return m_data; }
109 const Device& m_device;
113 template <
typename T> EIGEN_DEVICE_FUNC EIGEN_ALWAYS_INLINE
114 T loadConstant(
const T* address) {
118 #if defined(__CUDA_ARCH__) && __CUDA_ARCH__ >= 350 119 template <> EIGEN_DEVICE_FUNC EIGEN_ALWAYS_INLINE
120 float loadConstant(
const float* address) {
121 return __ldg(address);
123 template <> EIGEN_DEVICE_FUNC EIGEN_ALWAYS_INLINE
124 double loadConstant(
const double* address) {
125 return __ldg(address);
132 template<
typename Derived,
typename Device>
135 typedef typename Derived::Index Index;
136 typedef typename Derived::Scalar Scalar;
137 typedef typename Derived::Packet Packet;
138 typedef typename Derived::Scalar CoeffReturnType;
139 typedef typename Derived::Packet PacketReturnType;
140 typedef typename Derived::Dimensions Dimensions;
143 static const int NumCoords = internal::traits<Derived>::NumDimensions > 0 ?
144 internal::traits<Derived>::NumDimensions : 0;
147 IsAligned = Derived::IsAligned,
148 PacketAccess = Derived::PacketAccess,
149 Layout = Derived::Layout,
150 CoordAccess = NumCoords > 0,
153 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
TensorEvaluator(
const Derived& m,
const Device& device)
154 : m_data(m.data()), m_dims(m.dimensions()), m_device(device)
157 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
const Dimensions& dimensions()
const {
return m_dims; }
159 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
bool evalSubExprsIfNeeded(CoeffReturnType* data) {
160 if (!NumTraits<
typename internal::remove_const<Scalar>::type>::RequireInitialization && data) {
161 m_device.memcpy((
void*)data, m_data, m_dims.TotalSize() *
sizeof(Scalar));
167 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
void cleanup() { }
169 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE CoeffReturnType coeff(Index index)
const {
170 eigen_assert(m_data);
171 return loadConstant(m_data+index);
174 template<
int LoadMode> EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
175 PacketReturnType packet(Index index)
const 177 return internal::ploadt_ro<Packet, LoadMode>(m_data + index);
180 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE CoeffReturnType coeff(
const array<DenseIndex, NumCoords>& coords)
const {
181 eigen_assert(m_data);
182 const Index index = (
static_cast<int>(Layout) == static_cast<int>(ColMajor)) ? m_dims.IndexOfColMajor(coords)
183 : m_dims.IndexOfRowMajor(coords);
184 return loadConstant(m_data+index);
187 EIGEN_DEVICE_FUNC
const Scalar* data()
const {
return m_data; }
190 const Scalar* m_data;
192 const Device& m_device;
200 template<
typename NullaryOp,
typename ArgType,
typename Device>
201 struct TensorEvaluator<const TensorCwiseNullaryOp<NullaryOp, ArgType>, Device>
203 typedef TensorCwiseNullaryOp<NullaryOp, ArgType> XprType;
207 PacketAccess = internal::functor_traits<NullaryOp>::PacketAccess,
214 : m_functor(op.functor()), m_argImpl(op.nestedExpression(), device)
217 typedef typename XprType::Index Index;
218 typedef typename XprType::Scalar Scalar;
219 typedef typename internal::traits<XprType>::Scalar CoeffReturnType;
220 typedef typename internal::traits<XprType>::Packet PacketReturnType;
223 EIGEN_DEVICE_FUNC
const Dimensions& dimensions()
const {
return m_argImpl.dimensions(); }
225 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
bool evalSubExprsIfNeeded(CoeffReturnType*) {
return true; }
226 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
void cleanup() { }
228 EIGEN_DEVICE_FUNC CoeffReturnType coeff(Index index)
const 230 return m_functor(index);
233 template<
int LoadMode>
234 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE PacketReturnType packet(Index index)
const 236 return m_functor.template packetOp<Index,PacketReturnType>(index);
239 EIGEN_DEVICE_FUNC CoeffReturnType* data()
const {
return NULL; }
242 const NullaryOp m_functor;
250 template<
typename UnaryOp,
typename ArgType,
typename Device>
251 struct TensorEvaluator<const TensorCwiseUnaryOp<UnaryOp, ArgType>, Device>
253 typedef TensorCwiseUnaryOp<UnaryOp, ArgType> XprType;
262 EIGEN_DEVICE_FUNC
TensorEvaluator(
const XprType& op,
const Device& device)
263 : m_functor(op.functor()),
264 m_argImpl(op.nestedExpression(), device)
267 typedef typename XprType::Index Index;
268 typedef typename XprType::Scalar Scalar;
269 typedef typename internal::traits<XprType>::Scalar CoeffReturnType;
270 typedef typename internal::traits<XprType>::Packet PacketReturnType;
273 EIGEN_DEVICE_FUNC
const Dimensions& dimensions()
const {
return m_argImpl.dimensions(); }
275 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
bool evalSubExprsIfNeeded(Scalar*) {
276 m_argImpl.evalSubExprsIfNeeded(NULL);
279 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
void cleanup() {
283 EIGEN_DEVICE_FUNC CoeffReturnType coeff(Index index)
const 285 return m_functor(m_argImpl.coeff(index));
288 template<
int LoadMode>
289 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE PacketReturnType packet(Index index)
const 291 return m_functor.packetOp(m_argImpl.template packet<LoadMode>(index));
294 EIGEN_DEVICE_FUNC CoeffReturnType* data()
const {
return NULL; }
297 const UnaryOp m_functor;
304 template<
typename BinaryOp,
typename LeftArgType,
typename RightArgType,
typename Device>
305 struct TensorEvaluator<const TensorCwiseBinaryOp<BinaryOp, LeftArgType, RightArgType>, Device>
307 typedef TensorCwiseBinaryOp<BinaryOp, LeftArgType, RightArgType> XprType;
312 internal::functor_traits<BinaryOp>::PacketAccess,
317 EIGEN_DEVICE_FUNC
TensorEvaluator(
const XprType& op,
const Device& device)
318 : m_functor(op.functor()),
319 m_leftImpl(op.lhsExpression(), device),
320 m_rightImpl(op.rhsExpression(), device)
323 eigen_assert(dimensions_match(m_leftImpl.dimensions(), m_rightImpl.dimensions()));
326 typedef typename XprType::Index Index;
327 typedef typename XprType::Scalar Scalar;
328 typedef typename internal::traits<XprType>::Scalar CoeffReturnType;
329 typedef typename internal::traits<XprType>::Packet PacketReturnType;
332 EIGEN_DEVICE_FUNC
const Dimensions& dimensions()
const 335 return m_leftImpl.dimensions();
338 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
bool evalSubExprsIfNeeded(CoeffReturnType*) {
339 m_leftImpl.evalSubExprsIfNeeded(NULL);
340 m_rightImpl.evalSubExprsIfNeeded(NULL);
343 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
void cleanup() {
344 m_leftImpl.cleanup();
345 m_rightImpl.cleanup();
348 EIGEN_DEVICE_FUNC CoeffReturnType coeff(Index index)
const 350 return m_functor(m_leftImpl.coeff(index), m_rightImpl.coeff(index));
352 template<
int LoadMode>
353 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE PacketReturnType packet(Index index)
const 355 return m_functor.packetOp(m_leftImpl.template packet<LoadMode>(index), m_rightImpl.template packet<LoadMode>(index));
358 EIGEN_DEVICE_FUNC CoeffReturnType* data()
const {
return NULL; }
361 const BinaryOp m_functor;
369 template<
typename IfArgType,
typename ThenArgType,
typename ElseArgType,
typename Device>
370 struct TensorEvaluator<const TensorSelectOp<IfArgType, ThenArgType, ElseArgType>, Device>
372 typedef TensorSelectOp<IfArgType, ThenArgType, ElseArgType> XprType;
373 typedef typename XprType::Scalar Scalar;
378 internal::packet_traits<Scalar>::HasBlend,
383 EIGEN_DEVICE_FUNC
TensorEvaluator(
const XprType& op,
const Device& device)
384 : m_condImpl(op.ifExpression(), device),
385 m_thenImpl(op.thenExpression(), device),
386 m_elseImpl(op.elseExpression(), device)
390 eigen_assert(dimensions_match(m_condImpl.dimensions(), m_thenImpl.dimensions()));
391 eigen_assert(dimensions_match(m_thenImpl.dimensions(), m_elseImpl.dimensions()));
394 typedef typename XprType::Index Index;
395 typedef typename internal::traits<XprType>::Scalar CoeffReturnType;
396 typedef typename internal::traits<XprType>::Packet PacketReturnType;
399 EIGEN_DEVICE_FUNC
const Dimensions& dimensions()
const 402 return m_condImpl.dimensions();
405 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
bool evalSubExprsIfNeeded(CoeffReturnType*) {
406 m_condImpl.evalSubExprsIfNeeded(NULL);
407 m_thenImpl.evalSubExprsIfNeeded(NULL);
408 m_elseImpl.evalSubExprsIfNeeded(NULL);
411 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE
void cleanup() {
412 m_condImpl.cleanup();
413 m_thenImpl.cleanup();
414 m_elseImpl.cleanup();
417 EIGEN_DEVICE_FUNC CoeffReturnType coeff(Index index)
const 419 return m_condImpl.coeff(index) ? m_thenImpl.coeff(index) : m_elseImpl.coeff(index);
421 template<
int LoadMode>
422 EIGEN_DEVICE_FUNC PacketReturnType packet(Index index)
const 424 const int PacketSize = internal::unpacket_traits<PacketReturnType>::size;
425 internal::Selector<PacketSize> select;
426 for (Index i = 0; i < PacketSize; ++i) {
427 select.select[i] = m_condImpl.coeff(index+i);
429 return internal::pblend(select,
430 m_thenImpl.template packet<LoadMode>(index),
431 m_elseImpl.template packet<LoadMode>(index));
434 EIGEN_DEVICE_FUNC CoeffReturnType* data()
const {
return NULL; }
445 #endif // EIGEN_CXX11_TENSOR_TENSOR_EVALUATOR_H Namespace containing all symbols from the Eigen library.
Definition: CXX11Meta.h:13
The tensor evaluator classes.
Definition: TensorEvaluator.h:28