Builder for warp ids used to map scf.forall to reindexed threads. More...
#include "mlir/Dialect/GPU/TransformOps/Utils.h"
Inheritance diagram for mlir::transform::gpu::GpuThreadIdBuilder:Public Member Functions | |
| GpuThreadIdBuilder (MLIRContext *ctx, bool useLinearMapping=false, DeviceMaskingAttrInterface mask=nullptr) | |
| Public Member Functions inherited from mlir::transform::gpu::GpuIdBuilder | |
| GpuIdBuilder ()=default | |
| GpuIdBuilder (MLIRContext *ctx, bool useLinearMapping, const MappingIdBuilderFnType &builder) | |
Additional Inherited Members | |
| Public Types inherited from mlir::transform::gpu::GpuIdBuilder | |
| using | MappingIdBuilderFnType = std::function< DeviceMappingAttrInterface(MLIRContext *, mlir::gpu::MappingId)> |
| Public Attributes inherited from mlir::transform::gpu::GpuIdBuilder | |
| SmallVector< DeviceMappingAttrInterface > | mappingAttributes |
| The mapping attributes targeted by this generator. More... | |
| GpuIdBuilderFnType | idBuilder |
| The constructor that builds the concrete IR for mapping ids. More... | |
Detailed Description
Builder for warp ids used to map scf.forall to reindexed threads.
If useLinearMapping is false, the idBuilder method returns 3D values used for indexing rewrites as well as 3D sizes for predicate generation. If useLinearMapping is true, the idBuilder method returns nD values used for indexing rewrites as well as 1D sizes for predicate generation. If mask is provided, it will be used to filter the active threads.
Constructor & Destructor Documentation
◆ GpuThreadIdBuilder()
| mlir::transform::gpu::GpuThreadIdBuilder::GpuThreadIdBuilder | ( | MLIRContext * | ctx, |
| bool | useLinearMapping = false, |
||
| DeviceMaskingAttrInterface | mask = nullptr |
||
| ) |
Definition at line 338 of file Utils.cpp.
References mlir::get().
The documentation for this struct was generated from the following files:
