/home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-composable-kernel/checkouts/develop/include/ck/wrapper/utils/tensor_utils.hpp File Reference

/home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-composable-kernel/checkouts/develop/include/ck/wrapper/utils/tensor_utils.hpp File Reference#

Composable Kernel: /home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-composable-kernel/checkouts/develop/include/ck/wrapper/utils/tensor_utils.hpp File Reference
Typedefs | Functions
tensor_utils.hpp File Reference
#include "ck/ck.hpp"
#include "ck/utility/data_type.hpp"
#include "ck/utility/number.hpp"
#include "ck/utility/tuple.hpp"
#include "ck/utility/tuple_helper.hpp"
#include "ck/utility/dynamic_buffer.hpp"
#include "ck/utility/amd_address_space.hpp"
#include "ck/utility/multi_index.hpp"

Go to the source code of this file.

Typedefs

using MemoryTypeEnum = AddressSpaceEnum
 Memory type, allowed members: More...
 

Functions

template<MemoryTypeEnum MemoryType, typename ElementType , typename Shape , typename UnrolledDescriptorType >
constexpr auto make_tensor (ElementType *pointer, const Layout< Shape, UnrolledDescriptorType > &layout)
 Make tensor function. More...
 
template<MemoryTypeEnum MemoryType, typename ElementType , typename Shape , typename UnrolledDescriptorType >
constexpr auto make_register_tensor (const Layout< Shape, UnrolledDescriptorType > &layout)
 Make SGPR or VGPR tensor function. More...
 
template<MemoryTypeEnum BufferAddressSpace, typename ElementType , typename Shape , typename UnrolledDescriptorType >
__host__ __device__ void clear (Tensor< BufferAddressSpace, ElementType, Shape, UnrolledDescriptorType > &tensor)
 Clear tensor. (Only for Vpgr/Sgpr) More...
 
template<MemoryTypeEnum BufferAddressSpace, typename ElementType , typename Shape , typename UnrolledDescriptorType >
__host__ constexpr __device__ const auto & layout (const Tensor< BufferAddressSpace, ElementType, Shape, UnrolledDescriptorType > &tensor)
 Get Tensor Layout. More...
 
template<index_t... Idxs, MemoryTypeEnum BufferAddressSpace, typename ElementType , typename Shape , typename UnrolledDescriptorType >
__host__ constexpr __device__ auto size (const Tensor< BufferAddressSpace, ElementType, Shape, UnrolledDescriptorType > &tensor)
 Product of tensor shape dims. More...
 
template<index_t... Idxs, MemoryTypeEnum BufferAddressSpace, typename ElementType , typename Shape , typename UnrolledDescriptorType >
__host__ constexpr __device__ auto rank (const Tensor< BufferAddressSpace, ElementType, Shape, UnrolledDescriptorType > &tensor)
 Rank of Shape tuple. More...
 
template<index_t... Idxs, MemoryTypeEnum BufferAddressSpace, typename ElementType , typename Shape , typename UnrolledDescriptorType >
__host__ constexpr __device__ auto depth (const Tensor< BufferAddressSpace, ElementType, Shape, UnrolledDescriptorType > &tensor)
 Depth of Shape tuple. More...
 
template<MemoryTypeEnum BufferAddressSpace, typename ElementType , typename Shape , typename UnrolledDescriptorType >
__host__ constexpr __device__ const auto & shape (const Tensor< BufferAddressSpace, ElementType, Shape, UnrolledDescriptorType > &tensor)
 Get Tensor shape. More...
 
template<typename FromType , typename ToType >
constexpr auto slice (const FromType from, const ToType to)
 Get dim slice. More...
 
template<typename ToType >
constexpr auto slice (const ToType to)
 Get dim slice. (Assumed that from is equal to 1) More...
 
constexpr auto slice ()
 Get whole dim slice (from = 0, to = -1). More...
 

Typedef Documentation

◆ MemoryTypeEnum

using MemoryTypeEnum = AddressSpaceEnum

Memory type, allowed members:

  • Generic,
  • Global,
  • Lds,
  • Sgpr,
  • Vgpr,

Function Documentation

◆ clear()

template<MemoryTypeEnum BufferAddressSpace, typename ElementType , typename Shape , typename UnrolledDescriptorType >
__host__ __device__ void clear ( Tensor< BufferAddressSpace, ElementType, Shape, UnrolledDescriptorType > &  tensor)

Clear tensor. (Only for Vpgr/Sgpr)

Parameters
tensorTensor to be cleared.

◆ depth()

template<index_t... Idxs, MemoryTypeEnum BufferAddressSpace, typename ElementType , typename Shape , typename UnrolledDescriptorType >
__host__ constexpr __device__ auto depth ( const Tensor< BufferAddressSpace, ElementType, Shape, UnrolledDescriptorType > &  tensor)
constexpr

Depth of Shape tuple.

Template Parameters
IdxsIndexes to access specific shape dim (optional).
Parameters
tensorTensor to get depth of.
Returns
Requsted depth.

◆ layout()

template<MemoryTypeEnum BufferAddressSpace, typename ElementType , typename Shape , typename UnrolledDescriptorType >
__host__ constexpr __device__ const auto& layout ( const Tensor< BufferAddressSpace, ElementType, Shape, UnrolledDescriptorType > &  tensor)
constexpr

Get Tensor Layout.

Parameters
tensorTensor to get layout of.
Returns
Requsted layout.

◆ make_register_tensor()

template<MemoryTypeEnum MemoryType, typename ElementType , typename Shape , typename UnrolledDescriptorType >
constexpr auto make_register_tensor ( const Layout< Shape, UnrolledDescriptorType > &  layout)
constexpr

Make SGPR or VGPR tensor function.

Template Parameters
MemoryTypeType of memory.
ElementTypeMemory data type.
Returns
Constructed tensor.

◆ make_tensor()

template<MemoryTypeEnum MemoryType, typename ElementType , typename Shape , typename UnrolledDescriptorType >
constexpr auto make_tensor ( ElementType *  pointer,
const Layout< Shape, UnrolledDescriptorType > &  layout 
)
constexpr

Make tensor function.

Template Parameters
MemoryTypeType of memory.
Parameters
pointerPointer to the memory.
layoutTensor layout.
Returns
Constructed tensor.

◆ rank()

template<index_t... Idxs, MemoryTypeEnum BufferAddressSpace, typename ElementType , typename Shape , typename UnrolledDescriptorType >
__host__ constexpr __device__ auto rank ( const Tensor< BufferAddressSpace, ElementType, Shape, UnrolledDescriptorType > &  tensor)
constexpr

Rank of Shape tuple.

Template Parameters
IdxsIndexes to access specific shape dim (optional).
Parameters
tensorTensor to get rank of.
Returns
Requsted rank.

◆ shape()

template<MemoryTypeEnum BufferAddressSpace, typename ElementType , typename Shape , typename UnrolledDescriptorType >
__host__ constexpr __device__ const auto& shape ( const Tensor< BufferAddressSpace, ElementType, Shape, UnrolledDescriptorType > &  tensor)
constexpr

Get Tensor shape.

Parameters
tensorTensor to get shape from.
Returns
Requsted shape.

◆ size()

template<index_t... Idxs, MemoryTypeEnum BufferAddressSpace, typename ElementType , typename Shape , typename UnrolledDescriptorType >
__host__ constexpr __device__ auto size ( const Tensor< BufferAddressSpace, ElementType, Shape, UnrolledDescriptorType > &  tensor)
constexpr

Product of tensor shape dims.

Template Parameters
IdxsIndexes to access specific shape dim (optional).
Parameters
tensorTensor to get Shape of.
Returns
Requsted size.

◆ slice() [1/3]

constexpr auto slice ( )
constexpr

Get whole dim slice (from = 0, to = -1).

Returns
Requested slice. Could be used to create sliced tensor from other tensor.

◆ slice() [2/3]

template<typename FromType , typename ToType >
constexpr auto slice ( const FromType  from,
const ToType  to 
)
constexpr

Get dim slice.

Parameters
fromBeginning of the interval.
toEnd of the interval. (could be also negative to index from the end)
Returns
Requested slice. Could be used to create sliced tensor from other tensor.

◆ slice() [3/3]

template<typename ToType >
constexpr auto slice ( const ToType  to)
constexpr

Get dim slice. (Assumed that from is equal to 1)

Parameters
toEnd of the interval. (could be also negative to index from the end)
Returns
Requested slice. Could be used to create sliced tensor from other tensor.