tensor_coord.h
Go to the documentation of this file.
/***************************************************************************************************
* Copyright (c) 2017-2019, NVIDIA CORPORATION. All rights reserved.
*
* Redistribution and use in source and binary forms, with or without modification, are permitted
* provided that the following conditions are met:
* * Redistributions of source code must retain the above copyright notice, this list of
* conditions and the following disclaimer.
* * Redistributions in binary form must reproduce the above copyright notice, this list of
* conditions and the following disclaimer in the documentation and/or other materials
* provided with the distribution.
* * Neither the name of the NVIDIA CORPORATION nor the names of its contributors may be used
* to endorse or promote products derived from this software without specific prior written
* permission.
*
* THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" AND ANY EXPRESS OR
* IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND
* FITNESS FOR A PARTICULAR PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL NVIDIA CORPORATION BE LIABLE
* FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING,
* BUT NOT LIMITED TO, PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS;
* OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN CONTRACT,
* STRICT LIABILITY, OR TOR (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE
* OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
*
**************************************************************************************************/
#pragma once
#include "cutlass/cutlass.h"
#include "cutlass/coord.h"
namespace cutlass {
struct Tensor4DCoord : public Coord<4> {
using Base = Coord<4>;
using Index = typename Base::Index;
using LongIndex = typename Base::LongIndex;
static int const kN = 0;
static int const kH = 1;
static int const kW = 2;
static int const kC = 3;
//
// Methods
//
CUTLASS_HOST_DEVICE
Tensor4DCoord() { }
CUTLASS_HOST_DEVICE
Tensor4DCoord(Coord<4> const &coord): Base(coord) { }
CUTLASS_HOST_DEVICE
Tensor4DCoord(Index n, Index h, Index w, Index c): Base(make_Coord(n, h, w, c)) { }
CUTLASS_HOST_DEVICE
Index const & n() const { return this->at(kN); }
CUTLASS_HOST_DEVICE
Index & n() { return this->at(kN); }
CUTLASS_HOST_DEVICE
Index const & h() const { return this->at(kH); }
CUTLASS_HOST_DEVICE
Index & h() { return this->at(kH); }
CUTLASS_HOST_DEVICE
Index const & w() const { return this->at(kW); }
CUTLASS_HOST_DEVICE
Index & w() { return this->at(kW); }
CUTLASS_HOST_DEVICE
Index const & c() const { return this->at(kC); }
CUTLASS_HOST_DEVICE
Index & c() { return this->at(kC); }
//
// Coord operators
//
CUTLASS_HOST_DEVICE
Tensor4DCoord operator+(Base const& b) const {
return Tensor4DCoord(Base::operator+(b));
}
CUTLASS_HOST_DEVICE
Tensor4DCoord operator-(Base const& b) const {
return Tensor4DCoord(Base::operator-(b));
}
CUTLASS_HOST_DEVICE
Tensor4DCoord operator*(Base const& b) const {
return Tensor4DCoord(Base::operator*(b));
}
CUTLASS_HOST_DEVICE
Tensor4DCoord operator/(Base const& b) const {
return Tensor4DCoord(Base::operator/(b));
}
CUTLASS_HOST_DEVICE
Tensor4DCoord& operator+=(Base const& b) {
Base::operator+=(b);
return *this;
}
CUTLASS_HOST_DEVICE
Tensor4DCoord& operator-=(Base const& b) {
Base::operator-=(b);
return *this;
}
CUTLASS_HOST_DEVICE
Tensor4DCoord& operator*=(Base const& b) {
Base::operator*=(b);
return *this;
}
CUTLASS_HOST_DEVICE
Tensor4DCoord& operator/=(Base const& b) {
Base::operator/=(b);
return *this;
}
};
} // namespace cutlass
cutlass::Tensor4DCoord::n
CUTLASS_HOST_DEVICE Index & n()
Returns the batch of the coordinate.
Definition: tensor_coord.h:83
cutlass::Tensor4DCoord
Defines a canonical 4D coordinate used by tensor operations.
Definition: tensor_coord.h:38
cutlass
Definition: aligned_buffer.h:35
cutlass::Tensor4DCoord::operator/
CUTLASS_HOST_DEVICE Tensor4DCoord operator/(Base const &b) const
Element-wise division.
Definition: tensor_coord.h:133
cutlass::Coord< 4 >::operator*=
CUTLASS_HOST_DEVICE Coord & operator*=(Coord const &b)
In-place multiplication.
Definition: coord.h:222
coord.h
A Coord is a coordinate of arbitrary rank into a tensor or matrix.
cutlass::make_Coord
CUTLASS_HOST_DEVICE Coord< 1 > make_Coord(int _0)
Helper to make a 2-element coordinate.
Definition: coord.h:387
cutlass::Coord< 4 >::operator-=
CUTLASS_HOST_DEVICE Coord & operator-=(Coord const &b)
In-place subtraction.
Definition: coord.h:213
cutlass::Tensor4DCoord::w
CUTLASS_HOST_DEVICE Index & w()
Returns the column of the coordinate.
Definition: tensor_coord.h:99
cutlass::Tensor4DCoord::w
CUTLASS_HOST_DEVICE Index const & w() const
Returns the column of the coordinate.
Definition: tensor_coord.h:95
cutlass::Coord< 4 >::Index
int Index
Index type used to store elements.
Definition: coord.h:55
cutlass::Tensor4DCoord::operator*
CUTLASS_HOST_DEVICE Tensor4DCoord operator*(Base const &b) const
Element-wise multiplication.
Definition: tensor_coord.h:127
cutlass::Tensor4DCoord::Tensor4DCoord
CUTLASS_HOST_DEVICE Tensor4DCoord()
Default ctor.
Definition: tensor_coord.h:67
cutlass::Tensor4DCoord::c
CUTLASS_HOST_DEVICE Index const & c() const
Returns the channel of the coordinate.
Definition: tensor_coord.h:103
cutlass::Tensor4DCoord::kC
static int const kC
Channels dimension.
Definition: tensor_coord.h:59
cutlass::Tensor4DCoord::c
CUTLASS_HOST_DEVICE Index & c()
Returns the channel of the coordinate.
Definition: tensor_coord.h:107
CUTLASS_HOST_DEVICE
#define CUTLASS_HOST_DEVICE
Definition: cutlass.h:89
cutlass::Tensor4DCoord::LongIndex
typename Base::LongIndex LongIndex
LongIndex type.
Definition: tensor_coord.h:47
cutlass::Coord< 4 >::operator/=
CUTLASS_HOST_DEVICE Coord & operator/=(Coord const &b)
In-place division.
Definition: coord.h:231
cutlass::Tensor4DCoord::operator/=
CUTLASS_HOST_DEVICE Tensor4DCoord & operator/=(Base const &b)
In-place division.
Definition: tensor_coord.h:160
cutlass::Tensor4DCoord::operator-=
CUTLASS_HOST_DEVICE Tensor4DCoord & operator-=(Base const &b)
In-place subtraction.
Definition: tensor_coord.h:146
cutlass::Coord
Statically-sized array specifying Coords within a tensor.
Definition: coord.h:43
cutlass::Tensor4DCoord::n
CUTLASS_HOST_DEVICE Index const & n() const
Returns the batch of the coordinate.
Definition: tensor_coord.h:79
cutlass::Tensor4DCoord::Tensor4DCoord
CUTLASS_HOST_DEVICE Tensor4DCoord(Coord< 4 > const &coord)
Constructs from Coord<4>
Definition: tensor_coord.h:71
cutlass::Tensor4DCoord::h
CUTLASS_HOST_DEVICE Index & h()
Returns the row of the coordinate.
Definition: tensor_coord.h:91
cutlass::Tensor4DCoord::kN
static int const kN
Batch dimension.
Definition: tensor_coord.h:50
cutlass::Tensor4DCoord::Tensor4DCoord
CUTLASS_HOST_DEVICE Tensor4DCoord(Index n, Index h, Index w, Index c)
Helper to construct from N, H, W, and C.
Definition: tensor_coord.h:75
cutlass::Tensor4DCoord::kW
static int const kW
Width dimension.
Definition: tensor_coord.h:56
cutlass::Tensor4DCoord::operator+=
CUTLASS_HOST_DEVICE Tensor4DCoord & operator+=(Base const &b)
In-place addition.
Definition: tensor_coord.h:139
cutlass::Coord< 4 >::operator+=
CUTLASS_HOST_DEVICE Coord & operator+=(Coord const &b)
In-place addition.
Definition: coord.h:204
cutlass::Tensor4DCoord::h
CUTLASS_HOST_DEVICE Index const & h() const
Returns the row of the coordinate.
Definition: tensor_coord.h:87
cutlass::Coord< 4 >::LongIndex
int64_t LongIndex
Type used to represent linear offsets.
Definition: coord.h:58
cutlass::Tensor4DCoord::operator*=
CUTLASS_HOST_DEVICE Tensor4DCoord & operator*=(Base const &b)
In-place multiplication.
Definition: tensor_coord.h:153
cutlass::Coord< 4 >::at
CUTLASS_HOST_DEVICE Index & at()
Gets the index of a given Coord element.
Definition: coord.h:255
cutlass::Tensor4DCoord::operator+
CUTLASS_HOST_DEVICE Tensor4DCoord operator+(Base const &b) const
Element-wise addition.
Definition: tensor_coord.h:115
cutlass.h
Basic include for CUTLASS.
cutlass::Tensor4DCoord::kH
static int const kH
Height dimension.
Definition: tensor_coord.h:53
cutlass::Tensor4DCoord::Index
typename Base::Index Index
Index type.
Definition: tensor_coord.h:44
cutlass::Tensor4DCoord::operator-
CUTLASS_HOST_DEVICE Tensor4DCoord operator-(Base const &b) const
Element-wise subtraction.
Definition: tensor_coord.h:121