/home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-composable-kernel/checkouts/docs-7.0.0/include/ck_tile/core/arch/arch.hpp File Reference

/home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-composable-kernel/checkouts/docs-7.0.0/include/ck_tile/core/arch/arch.hpp File Reference#

Composable Kernel: /home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-composable-kernel/checkouts/docs-7.0.0/include/ck_tile/core/arch/arch.hpp File Reference
Classes | Namespaces | Macros | Typedefs | Enumerations | Functions
arch.hpp File Reference
#include "ck_tile/core/config.hpp"
#include "ck_tile/core/numeric/integer.hpp"
#include "ck_tile/core/numeric/integral_constant.hpp"

Go to the source code of this file.

Classes

struct  ck_tile::safe_underlying_type< T, true >
 
struct  ck_tile::safe_underlying_type< T, false >
 

Namespaces

 ck_tile
 

Macros

#define CK_CONSTANT_ADDRESS_SPACE
 

Typedefs

template<typename T >
using ck_tile::safe_underlying_type_t = typename safe_underlying_type< T, std::is_enum< T >::value >::type
 

Enumerations

enum class  ck_tile::address_space_enum : std::uint16_t {
  ck_tile::generic = 0 ,
  ck_tile::global ,
  ck_tile::lds ,
  ck_tile::sgpr ,
  ck_tile::constant ,
  ck_tile::vgpr
}
 
enum class  ck_tile::memory_operation_enum : std::uint16_t {
  ck_tile::set = 0 ,
  ck_tile::atomic_add ,
  ck_tile::atomic_max ,
  ck_tile::add
}
 

Functions

constexpr CK_TILE_HOST_DEVICE index_t ck_tile::get_warp_size ()
 
CK_TILE_DEVICE index_t ck_tile::get_grid_size ()
 
CK_TILE_DEVICE index_t ck_tile::get_block_size ()
 
CK_TILE_DEVICE index_t ck_tile::get_thread_local_1d_id ()
 
CK_TILE_DEVICE index_t ck_tile::get_thread_global_1d_id ()
 
CK_TILE_DEVICE index_t ck_tile::get_block_1d_id ()
 
CK_TILE_DEVICE index_t ck_tile::get_lane_id ()
 
CK_TILE_DEVICE index_t ck_tile::get_warp_id ()
 
CK_TILE_DEVICE index_t ck_tile::get_thread_id ()
 
CK_TILE_DEVICE index_t ck_tile::get_block_id ()
 
CK_TILE_DEVICE void ck_tile::block_sync_lds ()
 
CK_TILE_DEVICE void ck_tile::block_sync_load_raw (index_t cnt=0)
 
CK_TILE_DEVICE void ck_tile::block_sync_lds_direct_load ()
 
CK_TILE_DEVICE void ck_tile::s_nop (index_t cnt=0)
 
template<typename T >
__device__ T * ck_tile::cast_pointer_to_generic_address_space (T CK_CONSTANT_ADDRESS_SPACE *p)
 
template<typename T >
__host__ __device__ T CK_CONSTANT_ADDRESS_SPACE * ck_tile::cast_pointer_to_constant_address_space (T *p)
 
constexpr CK_TILE_HOST_DEVICE index_t ck_tile::get_smem_capacity ()
 

Macro Definition Documentation

◆ CK_CONSTANT_ADDRESS_SPACE

#define CK_CONSTANT_ADDRESS_SPACE
Value:
__attribute__((address_space( \
static_cast<safe_underlying_type_t<address_space_enum>>(address_space_enum::constant))))