9 #ifndef MLIR_DIALECT_GPU_TRANSFORMOPS_UTILS_H
10 #define MLIR_DIALECT_GPU_TRANSFORMOPS_UTILS_H
22 enum class MappingId : uint64_t;
42 llvm::interleaveComma(res.
mappingIdOps, os <<
"----mappingIdOps: ");
44 llvm::interleaveComma(res.
predicateOps, os <<
"----predicateOps: ");
83 DeviceMaskingAttrInterface mask =
nullptr);
94 bool useLinearMapping =
false,
95 DeviceMaskingAttrInterface mask =
nullptr);
96 int64_t warpSize = 32;
98 static constexpr int64_t kNumWarpsPerGroup = 4;
109 bool useLinearMapping =
false,
110 DeviceMaskingAttrInterface mask =
nullptr);
111 int64_t warpSize = 32;
122 DeviceMaskingAttrInterface mask =
nullptr);
132 DeviceMaskingAttrInterface mask =
nullptr);
133 int64_t warpSize = 32;
140 std::optional<int64_t> gridDimX,
141 std::optional<int64_t> gridDimY,
142 std::optional<int64_t> gridDimZ,
143 std::optional<int64_t> blockDimX,
144 std::optional<int64_t> blockDimY,
145 std::optional<int64_t> blockDimZ);
151 TransformOpInterface transformOp, mlir::gpu::LaunchOp &launchOp,
152 std::optional<int64_t> gridDimX = std::nullopt,
153 std::optional<int64_t> gridDimY = std::nullopt,
154 std::optional<int64_t> gridDimZ = std::nullopt,
155 std::optional<int64_t> blockDimX = std::nullopt,
156 std::optional<int64_t> blockDimY = std::nullopt,
157 std::optional<int64_t> blockDimZ = std::nullopt);
162 TransformOpInterface transformOp,
163 std::optional<int64_t> gridDimX = std::nullopt,
164 std::optional<int64_t> gridDimY = std::nullopt,
165 std::optional<int64_t> gridDimZ = std::nullopt,
166 std::optional<int64_t> blockDimX = std::nullopt,
167 std::optional<int64_t> blockDimY = std::nullopt,
168 std::optional<int64_t> blockDimZ = std::nullopt);
173 TransformOpInterface transformOp);
The result of a transform IR operation application.
This class defines the main interface for locations in MLIR and acts as a non-nullable wrapper around...
MLIRContext is the top-level object for a collection of MLIR operations.
Operation is the basic unit of execution within MLIR.
This class coordinates the application of a rewrite on a set of IR, providing a way for clients to tr...
Include the generated interface declarations.
raw_ostream & operator<<(raw_ostream &os, const AliasResult &result)