Classes
class	SerializeGPUModuleBase
	Base class for all ROCDL serializations from GPU modules into binary strings. More...

Enumerations
enum class	AMDGCNLibraries : uint32_t { None = 0 , Ockl = 1 , Ocml = 2 , OpenCL = 4 , Hip = 8 , LastLib = Hip , LLVM_MARK_AS_BITMASK_ENUM =(LastLib) , All = (LastLib << 1) - 1 }
	Helper enum for specifying the AMD GCN device libraries required for compilation. More...

Functions
void	registerROCDLTargetInterfaceExternalModels (DialectRegistry &registry)
	Registers the `TargetAttrInterface` for the `#rocdl.target` attribute in the given registry. More...

void	registerROCDLTargetInterfaceExternalModels (MLIRContext &context)
	Registers the `TargetAttrInterface` for the `#rocdl.target` attribute in the registry associated with the given context. More...

StringRef	getROCMPath ()
	Searches & returns the path ROCM toolkit path, the search order is: More...

std::optional< DenseMap< StringAttr, NamedAttrList > >	getAMDHSAKernelsELFMetadata (Builder &builder, ArrayRef< char > elfData)
	Returns a map containing the `amdhsa.kernels` ELF metadata for each of the kernels in the binary, or `std::nullopt` if the metadata couldn't be retrieved. More...

gpu::KernelTableAttr	getKernelMetadata (Operation *gpuModule, ArrayRef< char > elfData={})
	Returns a `#gpu.kernel_table` containing kernel metadata for each of the kernels in `gpuModule`. More...

Enumeration Type Documentation

◆ AMDGCNLibraries

enum mlir::ROCDL::AMDGCNLibraries : uint32_t

strong

Helper enum for specifying the AMD GCN device libraries required for compilation.

Enumerator
None
Ockl
Ocml
OpenCL
Hip
LastLib
LLVM_MARK_AS_BITMASK_ENUM
All

Definition at line 34 of file Utils.h.

Function Documentation

◆ getAMDHSAKernelsELFMetadata()

std::optional< DenseMap< StringAttr, NamedAttrList > > mlir::ROCDL::getAMDHSAKernelsELFMetadata	(	Builder &	builder,
		ArrayRef< char >	elfData
	)

Returns a map containing the amdhsa.kernels ELF metadata for each of the kernels in the binary, or std::nullopt if the metadata couldn't be retrieved.

The map associates the name of the kernel with the list of named attributes found in amdhsa.kernels. For more information on the ELF metadata see: https://llvm.org/docs/AMDGPUUsage.html#amdhsa

Definition at line 24 of file Utils.cpp.

References mlir::NamedAttrList::append(), mlir::Builder::getDenseI32ArrayAttr(), mlir::Builder::getI64IntegerAttr(), and mlir::Builder::getStringAttr().

Referenced by getKernelMetadata().

◆ getKernelMetadata()

gpu::KernelTableAttr mlir::ROCDL::getKernelMetadata	(	Operation *	gpuModule,
		ArrayRef< char >	elfData = `{}`
	)

Returns a #gpu.kernel_table containing kernel metadata for each of the kernels in gpuModule.

If elfData is valid, then the amdhsa.kernels ELF metadata will be added to the #gpu.kernel_table.

Definition at line 71 of file Utils.cpp.

References mlir::get(), getAMDHSAKernelsELFMetadata(), mlir::Operation::getContext(), and mlir::Builder::getDictionaryAttr().

◆ getROCMPath()

StringRef mlir::ROCDL::getROCMPath ( )

Searches & returns the path ROCM toolkit path, the search order is:

The ROCM_PATH environment variable.
The ROCM_ROOT environment variable.
The ROCM_HOME environment variable.
The ROCM path detected by CMake.
Returns an empty string.

Definition at line 84 of file Target.cpp.

References __DEFAULT_ROCM_PATH__.

Referenced by mlir::ROCDL::SerializeGPUModuleBase::SerializeGPUModuleBase().

◆ registerROCDLTargetInterfaceExternalModels() [1/2]

void mlir::ROCDL::registerROCDLTargetInterfaceExternalModels ( DialectRegistry & registry )

Registers the TargetAttrInterface for the #rocdl.target attribute in the given registry.

Definition at line 69 of file Target.cpp.

References mlir::DialectRegistry::addExtension().

Referenced by mlir::registerAllDialects(), and registerROCDLTargetInterfaceExternalModels().

◆ registerROCDLTargetInterfaceExternalModels() [2/2]

void mlir::ROCDL::registerROCDLTargetInterfaceExternalModels ( MLIRContext & context )

Registers the TargetAttrInterface for the #rocdl.target attribute in the registry associated with the given context.

Definition at line 76 of file Target.cpp.

References mlir::MLIRContext::appendDialectRegistry(), and registerROCDLTargetInterfaceExternalModels().

Classes

Enumerations

Functions