PoolKernel< Problem_, Policy_ > Struct Template Reference

• ck_tile
• PoolKernel

Public Types | Public Member Functions | Static Public Member Functions | Static Public Attributes | List of all members

ck_tile::PoolKernel< Problem_, Policy_ > Struct Template Reference

#include <pool_kernel.hpp>

Public Types
using	Problem = ck_tile::remove_cvref_t<Problem_>
using	Policy = ck_tile::remove_cvref_t<Policy_>
using	InDataType = ck_tile::remove_cvref_t<typename Problem::InDataType>
using	ComputeDataType = ck_tile::remove_cvref_t<typename Problem::ComputeDataType>
using	OutDataType = ck_tile::remove_cvref_t<typename Problem::OutDataType>
using	IndexDataType = ck_tile::remove_cvref_t<typename Problem::IndexDataType>

Public Member Functions
template<typename TensorShape, typename WindowShape>
CK_TILE_DEVICE void	operator() (PoolKernelArgs< TensorShape, WindowShape > kargs) const

Static Public Member Functions
static CK_TILE_HOST constexpr auto	BlockSize ()
template<typename TensorShape, typename WindowShape>
static CK_TILE_DEVICE auto	MakeTensorView2D (PoolKernelArgs< TensorShape, WindowShape > kargs)
template<typename TensorShape, typename WindowShape>
static CK_TILE_DEVICE auto	MakeTensorView3D (PoolKernelArgs< TensorShape, WindowShape > kargs)
template<typename TensorShape, typename WindowShape>
static CK_TILE_HOST bool	IsSupportedArgument (PoolKernelArgs< TensorShape, WindowShape > kargs)
	Validates if the given arguments are supported by the pooling kernel.
template<typename TensorShape, typename WindowShape>
static CK_TILE_HOST constexpr index_t	CalculateGridSize (PoolKernelArgs< TensorShape, WindowShape > kargs)
template<typename TensorShape, typename WindowShape>
static CK_TILE_HOST constexpr auto	MakeKernelArgs (PoolHostArgs< TensorShape, WindowShape > &host_args)
	Create kernel arguments from host arguments.

Static Public Attributes
static constexpr index_t	kBlockSize = Problem::BlockShape::BlockSize

Member Typedef Documentation

◆ ComputeDataType

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

using ck_tile::PoolKernel< Problem_, Policy_ >::ComputeDataType = ck_tile::remove_cvref_t<typename Problem::ComputeDataType>

◆ InDataType

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

using ck_tile::PoolKernel< Problem_, Policy_ >::InDataType = ck_tile::remove_cvref_t<typename Problem::InDataType>

◆ IndexDataType

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

using ck_tile::PoolKernel< Problem_, Policy_ >::IndexDataType = ck_tile::remove_cvref_t<typename Problem::IndexDataType>

◆ OutDataType

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

using ck_tile::PoolKernel< Problem_, Policy_ >::OutDataType = ck_tile::remove_cvref_t<typename Problem::OutDataType>

◆ Policy

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

using ck_tile::PoolKernel< Problem_, Policy_ >::Policy = ck_tile::remove_cvref_t<Policy_>

◆ Problem

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

using ck_tile::PoolKernel< Problem_, Policy_ >::Problem = ck_tile::remove_cvref_t<Problem_>

Member Function Documentation

◆ BlockSize()

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

CK_TILE_HOST constexpr auto ck_tile::PoolKernel< Problem_, Policy_ >::BlockSize ( )

inlinestaticconstexpr

◆ CalculateGridSize()

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

template<typename TensorShape, typename WindowShape>

CK_TILE_HOST constexpr index_t ck_tile::PoolKernel< Problem_, Policy_ >::CalculateGridSize ( PoolKernelArgs< TensorShape, WindowShape > kargs )

inlinestaticconstexpr

Parameters

kargs

The pooling kernel arguments

Returns

The calculated grid size

◆ IsSupportedArgument()

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

template<typename TensorShape, typename WindowShape>

CK_TILE_HOST bool ck_tile::PoolKernel< Problem_, Policy_ >::IsSupportedArgument ( PoolKernelArgs< TensorShape, WindowShape > kargs )

inlinestatic

Validates if the given arguments are supported by the pooling kernel.

Parameters

kargs

The pooling kernel arguments containing all necessary parameters.

Returns

true if the arguments are supported, false otherwise.

Note

Requirements:

• Last dimension (C) must be contiguous (stride = 1) for vectorized access
• Window dimensions must be supported (2D or 3D)
• All dimension sizes must be consistent between input and output

◆ MakeKernelArgs()

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

template<typename TensorShape, typename WindowShape>

CK_TILE_HOST constexpr auto ck_tile::PoolKernel< Problem_, Policy_ >::MakeKernelArgs ( PoolHostArgs< TensorShape, WindowShape > & host_args )

inlinestaticconstexpr

Create kernel arguments from host arguments.

◆ MakeTensorView2D()

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

template<typename TensorShape, typename WindowShape>

CK_TILE_DEVICE auto ck_tile::PoolKernel< Problem_, Policy_ >::MakeTensorView2D ( PoolKernelArgs< TensorShape, WindowShape > kargs )

inlinestatic

◆ MakeTensorView3D()

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

template<typename TensorShape, typename WindowShape>

CK_TILE_DEVICE auto ck_tile::PoolKernel< Problem_, Policy_ >::MakeTensorView3D ( PoolKernelArgs< TensorShape, WindowShape > kargs )

inlinestatic

◆ operator()()

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

template<typename TensorShape, typename WindowShape>

CK_TILE_DEVICE void ck_tile::PoolKernel< Problem_, Policy_ >::operator() ( PoolKernelArgs< TensorShape, WindowShape > kargs ) const

inline

Member Data Documentation

◆ kBlockSize

template<typename Problem_, typename Policy_ = PoolDefaultPolicy>

index_t ck_tile::PoolKernel< Problem_, Policy_ >::kBlockSize = Problem::BlockShape::BlockSize

staticconstexpr

---

The documentation for this struct was generated from the following file:

• pool_kernel.hpp