SWDEV-484188 - Moving std::maps into struct const and into amd::Kernel class.
Change-Id: Ie4d5a64511412fdb498b045aaffb52c3a1286de6
This commit is contained in:
kjayapra-amd
committed by
Karthik Jayaprakash
Karthik Jayaprakash
parent
4d0b815d06
commit
4ecd77df5e
+11
-171
@@ -33,7 +33,7 @@ class KernelSignature;
|
||||
class NDRange;
|
||||
|
||||
struct KernelParameterDescriptor {
|
||||
enum {
|
||||
enum Desc {
|
||||
Value = 0,
|
||||
MemoryObject = 1,
|
||||
ReferenceObject = 2,
|
||||
@@ -65,7 +65,8 @@ struct KernelParameterDescriptor {
|
||||
HiddenSharedBase = 28,
|
||||
HiddenQueuePtr = 29,
|
||||
HiddenDynamicLdsSize = 30,
|
||||
HiddenLast = 31
|
||||
HiddenLast = 31,
|
||||
MaxSize = 32,
|
||||
};
|
||||
clk_value_type_t type_; //!< The parameter's type
|
||||
size_t offset_; //!< Its offset in the parameter's stack
|
||||
@@ -122,14 +123,16 @@ enum class ArgField : uint8_t {
|
||||
IsRestrict = 10,
|
||||
IsVolatile = 11,
|
||||
IsPipe = 12,
|
||||
Offset = 13
|
||||
Offset = 13,
|
||||
MaxSize = 14
|
||||
};
|
||||
|
||||
enum class AttrField : uint8_t {
|
||||
ReqdWorkGroupSize = 0,
|
||||
WorkGroupSizeHint = 1,
|
||||
VecTypeHint = 2,
|
||||
RuntimeHandle = 3
|
||||
RuntimeHandle = 3,
|
||||
MaxSize = 4,
|
||||
};
|
||||
|
||||
enum class CodePropField : uint8_t {
|
||||
@@ -144,84 +147,8 @@ enum class CodePropField : uint8_t {
|
||||
IsDynamicCallStack = 8,
|
||||
IsXNACKEnabled = 9,
|
||||
NumSpilledSGPRs = 10,
|
||||
NumSpilledVGPRs = 11
|
||||
};
|
||||
|
||||
|
||||
static const std::map<std::string, ArgField> ArgFieldMap =
|
||||
{
|
||||
{"Name", ArgField::Name},
|
||||
{"TypeName", ArgField::TypeName},
|
||||
{"Size", ArgField::Size},
|
||||
{"Align", ArgField::Align},
|
||||
{"ValueKind", ArgField::ValueKind},
|
||||
{"PointeeAlign", ArgField::PointeeAlign},
|
||||
{"AddrSpaceQual", ArgField::AddrSpaceQual},
|
||||
{"AccQual", ArgField::AccQual},
|
||||
{"ActualAccQual", ArgField::ActualAccQual},
|
||||
{"IsConst", ArgField::IsConst},
|
||||
{"IsRestrict", ArgField::IsRestrict},
|
||||
{"IsVolatile", ArgField::IsVolatile},
|
||||
{"IsPipe", ArgField::IsPipe}
|
||||
};
|
||||
|
||||
static const std::map<std::string, uint32_t> ArgValueKind = {
|
||||
{"ByValue", amd::KernelParameterDescriptor::ValueObject},
|
||||
{"GlobalBuffer", amd::KernelParameterDescriptor::MemoryObject},
|
||||
{"DynamicSharedPointer", amd::KernelParameterDescriptor::MemoryObject},
|
||||
{"Sampler", amd::KernelParameterDescriptor::SamplerObject},
|
||||
{"Image", amd::KernelParameterDescriptor::ImageObject },
|
||||
{"Pipe", amd::KernelParameterDescriptor::MemoryObject},
|
||||
{"Queue", amd::KernelParameterDescriptor::QueueObject},
|
||||
{"HiddenGlobalOffsetX", amd::KernelParameterDescriptor::HiddenGlobalOffsetX},
|
||||
{"HiddenGlobalOffsetY", amd::KernelParameterDescriptor::HiddenGlobalOffsetY},
|
||||
{"HiddenGlobalOffsetZ", amd::KernelParameterDescriptor::HiddenGlobalOffsetZ},
|
||||
{"HiddenNone", amd::KernelParameterDescriptor::HiddenNone},
|
||||
{"HiddenPrintfBuffer", amd::KernelParameterDescriptor::HiddenPrintfBuffer},
|
||||
{"HiddenDefaultQueue", amd::KernelParameterDescriptor::HiddenDefaultQueue},
|
||||
{"HiddenCompletionAction", amd::KernelParameterDescriptor::HiddenCompletionAction},
|
||||
{"HiddenMultigridSyncArg", amd::KernelParameterDescriptor::HiddenMultiGridSync},
|
||||
{"HiddenHostcallBuffer", amd::KernelParameterDescriptor::HiddenHostcallBuffer}
|
||||
};
|
||||
|
||||
static const std::map<std::string, cl_kernel_arg_access_qualifier> ArgAccQual = {
|
||||
{"Default", CL_KERNEL_ARG_ACCESS_NONE},
|
||||
{"ReadOnly", CL_KERNEL_ARG_ACCESS_READ_ONLY},
|
||||
{"WriteOnly", CL_KERNEL_ARG_ACCESS_WRITE_ONLY},
|
||||
{"ReadWrite", CL_KERNEL_ARG_ACCESS_READ_WRITE}
|
||||
};
|
||||
|
||||
static const std::map<std::string, cl_kernel_arg_address_qualifier> ArgAddrSpaceQual = {
|
||||
{"Private", CL_KERNEL_ARG_ADDRESS_PRIVATE},
|
||||
{"Global", CL_KERNEL_ARG_ADDRESS_GLOBAL},
|
||||
{"Constant", CL_KERNEL_ARG_ADDRESS_CONSTANT},
|
||||
{"Local", CL_KERNEL_ARG_ADDRESS_LOCAL},
|
||||
{"Generic", CL_KERNEL_ARG_ADDRESS_GLOBAL},
|
||||
{"Region", CL_KERNEL_ARG_ADDRESS_PRIVATE}
|
||||
};
|
||||
|
||||
static const std::map<std::string,AttrField> AttrFieldMap =
|
||||
{
|
||||
{"ReqdWorkGroupSize", AttrField::ReqdWorkGroupSize},
|
||||
{"WorkGroupSizeHint", AttrField::WorkGroupSizeHint},
|
||||
{"VecTypeHint", AttrField::VecTypeHint},
|
||||
{"RuntimeHandle", AttrField::RuntimeHandle}
|
||||
};
|
||||
|
||||
static const std::map<std::string,CodePropField> CodePropFieldMap =
|
||||
{
|
||||
{"KernargSegmentSize", CodePropField::KernargSegmentSize},
|
||||
{"GroupSegmentFixedSize", CodePropField::GroupSegmentFixedSize},
|
||||
{"PrivateSegmentFixedSize", CodePropField::PrivateSegmentFixedSize},
|
||||
{"KernargSegmentAlign", CodePropField::KernargSegmentAlign},
|
||||
{"WavefrontSize", CodePropField::WavefrontSize},
|
||||
{"NumSGPRs", CodePropField::NumSGPRs},
|
||||
{"NumVGPRs", CodePropField::NumVGPRs},
|
||||
{"MaxFlatWorkGroupSize", CodePropField::MaxFlatWorkGroupSize},
|
||||
{"IsDynamicCallStack", CodePropField::IsDynamicCallStack},
|
||||
{"IsXNACKEnabled", CodePropField::IsXNACKEnabled},
|
||||
{"NumSpilledSGPRs", CodePropField::NumSpilledSGPRs},
|
||||
{"NumSpilledVGPRs", CodePropField::NumSpilledVGPRs}
|
||||
NumSpilledVGPRs = 11,
|
||||
MaxSize = 12,
|
||||
};
|
||||
|
||||
// for Code Object V3
|
||||
@@ -243,95 +170,8 @@ enum class KernelField : uint8_t {
|
||||
NumSpilledVGPRs = 14,
|
||||
Kind = 15,
|
||||
WgpMode = 16,
|
||||
UniformWrokGroupSize = 17
|
||||
};
|
||||
|
||||
static const std::map<std::string,ArgField> ArgFieldMapV3 =
|
||||
{
|
||||
{".name", ArgField::Name},
|
||||
{".type_name", ArgField::TypeName},
|
||||
{".size", ArgField::Size},
|
||||
{".offset", ArgField::Offset},
|
||||
{".value_kind", ArgField::ValueKind},
|
||||
{".pointee_align", ArgField::PointeeAlign},
|
||||
{".address_space", ArgField::AddrSpaceQual},
|
||||
{".access", ArgField::AccQual},
|
||||
{".actual_access", ArgField::ActualAccQual},
|
||||
{".is_const", ArgField::IsConst},
|
||||
{".is_restrict", ArgField::IsRestrict},
|
||||
{".is_volatile", ArgField::IsVolatile},
|
||||
{".is_pipe", ArgField::IsPipe}
|
||||
};
|
||||
|
||||
static const std::map<std::string, uint32_t> ArgValueKindV3 = {
|
||||
{"by_value", amd::KernelParameterDescriptor::ValueObject},
|
||||
{"global_buffer", amd::KernelParameterDescriptor::MemoryObject},
|
||||
{"dynamic_shared_pointer", amd::KernelParameterDescriptor::MemoryObject},
|
||||
{"sampler", amd::KernelParameterDescriptor::SamplerObject},
|
||||
{"image", amd::KernelParameterDescriptor::ImageObject },
|
||||
{"pipe", amd::KernelParameterDescriptor::MemoryObject},
|
||||
{"queue", amd::KernelParameterDescriptor::QueueObject},
|
||||
{"hidden_global_offset_x", amd::KernelParameterDescriptor::HiddenGlobalOffsetX},
|
||||
{"hidden_global_offset_y", amd::KernelParameterDescriptor::HiddenGlobalOffsetY},
|
||||
{"hidden_global_offset_z", amd::KernelParameterDescriptor::HiddenGlobalOffsetZ},
|
||||
{"hidden_none", amd::KernelParameterDescriptor::HiddenNone},
|
||||
{"hidden_printf_buffer", amd::KernelParameterDescriptor::HiddenPrintfBuffer},
|
||||
{"hidden_default_queue", amd::KernelParameterDescriptor::HiddenDefaultQueue},
|
||||
{"hidden_completion_action", amd::KernelParameterDescriptor::HiddenCompletionAction},
|
||||
{"hidden_multigrid_sync_arg", amd::KernelParameterDescriptor::HiddenMultiGridSync},
|
||||
{"hidden_heap_v1", amd::KernelParameterDescriptor::HiddenHeap},
|
||||
{"hidden_hostcall_buffer", amd::KernelParameterDescriptor::HiddenHostcallBuffer},
|
||||
{"hidden_block_count_x", amd::KernelParameterDescriptor::HiddenBlockCountX},
|
||||
{"hidden_block_count_y", amd::KernelParameterDescriptor::HiddenBlockCountY},
|
||||
{"hidden_block_count_z", amd::KernelParameterDescriptor::HiddenBlockCountZ},
|
||||
{"hidden_group_size_x", amd::KernelParameterDescriptor::HiddenGroupSizeX},
|
||||
{"hidden_group_size_y", amd::KernelParameterDescriptor::HiddenGroupSizeY},
|
||||
{"hidden_group_size_z", amd::KernelParameterDescriptor::HiddenGroupSizeZ},
|
||||
{"hidden_remainder_x", amd::KernelParameterDescriptor::HiddenRemainderX},
|
||||
{"hidden_remainder_y", amd::KernelParameterDescriptor::HiddenRemainderY},
|
||||
{"hidden_remainder_z", amd::KernelParameterDescriptor::HiddenRemainderZ},
|
||||
{"hidden_grid_dims", amd::KernelParameterDescriptor::HiddenGridDims},
|
||||
{"hidden_private_base", amd::KernelParameterDescriptor::HiddenPrivateBase},
|
||||
{"hidden_shared_base", amd::KernelParameterDescriptor::HiddenSharedBase},
|
||||
{"hidden_queue_ptr", amd::KernelParameterDescriptor::HiddenQueuePtr},
|
||||
{"hidden_dynamic_lds_size", amd::KernelParameterDescriptor::HiddenDynamicLdsSize}
|
||||
};
|
||||
|
||||
static const std::map<std::string, cl_kernel_arg_access_qualifier> ArgAccQualV3 = {
|
||||
{"default", CL_KERNEL_ARG_ACCESS_NONE},
|
||||
{"read_only", CL_KERNEL_ARG_ACCESS_READ_ONLY},
|
||||
{"write_only", CL_KERNEL_ARG_ACCESS_WRITE_ONLY},
|
||||
{"read_write", CL_KERNEL_ARG_ACCESS_READ_WRITE}
|
||||
};
|
||||
|
||||
static const std::map<std::string, cl_kernel_arg_address_qualifier> ArgAddrSpaceQualV3 = {
|
||||
{"private", CL_KERNEL_ARG_ADDRESS_PRIVATE},
|
||||
{"global", CL_KERNEL_ARG_ADDRESS_GLOBAL},
|
||||
{"constant", CL_KERNEL_ARG_ADDRESS_CONSTANT},
|
||||
{"local", CL_KERNEL_ARG_ADDRESS_LOCAL},
|
||||
{"generic", CL_KERNEL_ARG_ADDRESS_GLOBAL},
|
||||
{"region", CL_KERNEL_ARG_ADDRESS_PRIVATE}
|
||||
};
|
||||
|
||||
static const std::map<std::string,KernelField> KernelFieldMapV3 = {
|
||||
{".symbol", KernelField::SymbolName},
|
||||
{".reqd_workgroup_size", KernelField::ReqdWorkGroupSize},
|
||||
{".workgroup_size_hint", KernelField::WorkGroupSizeHint},
|
||||
{".vec_type_hint", KernelField::VecTypeHint},
|
||||
{".device_enqueue_symbol", KernelField::DeviceEnqueueSymbol},
|
||||
{".kernarg_segment_size", KernelField::KernargSegmentSize},
|
||||
{".group_segment_fixed_size", KernelField::GroupSegmentFixedSize},
|
||||
{".private_segment_fixed_size", KernelField::PrivateSegmentFixedSize},
|
||||
{".kernarg_segment_align", KernelField::KernargSegmentAlign},
|
||||
{".wavefront_size", KernelField::WavefrontSize},
|
||||
{".sgpr_count", KernelField::NumSGPRs},
|
||||
{".vgpr_count", KernelField::NumVGPRs},
|
||||
{".max_flat_workgroup_size", KernelField::MaxFlatWorkGroupSize},
|
||||
{".sgpr_spill_count", KernelField::NumSpilledSGPRs},
|
||||
{".vgpr_spill_count", KernelField::NumSpilledVGPRs},
|
||||
{".kind", KernelField::Kind},
|
||||
{".workgroup_processor_mode", KernelField::WgpMode},
|
||||
{".uniform_work_group_size", KernelField::UniformWrokGroupSize}
|
||||
UniformWrokGroupSize = 17,
|
||||
MaxSize = 18
|
||||
};
|
||||
|
||||
#endif // defined(USE_COMGR_LIBRARY)
|
||||
|
||||
Reference in New Issue
Block a user