Bug Summary

File:build/source/build-llvm/tools/clang/stage2-bins/tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc
Warning:line 2830, column 7
Value stored to '_firstPrinted' is never read

Annotated Source Code

Press '?' to see keyboard shortcuts

clang -cc1 -cc1 -triple x86_64-pc-linux-gnu -analyze -disable-free -clear-ast-before-backend -disable-llvm-verifier -discard-value-names -main-file-name SPIRVAttributes.cpp -analyzer-checker=core -analyzer-checker=apiModeling -analyzer-checker=unix -analyzer-checker=deadcode -analyzer-checker=cplusplus -analyzer-checker=security.insecureAPI.UncheckedReturn -analyzer-checker=security.insecureAPI.getpw -analyzer-checker=security.insecureAPI.gets -analyzer-checker=security.insecureAPI.mktemp -analyzer-checker=security.insecureAPI.mkstemp -analyzer-checker=security.insecureAPI.vfork -analyzer-checker=nullability.NullPassedToNonnull -analyzer-checker=nullability.NullReturnedFromNonnull -analyzer-output plist -w -setup-static-analyzer -analyzer-config-compatibility-mode=true -mrelocation-model pic -pic-level 2 -mframe-pointer=none -fmath-errno -ffp-contract=on -fno-rounding-math -mconstructor-aliases -funwind-tables=2 -target-cpu x86-64 -tune-cpu generic -debugger-tuning=gdb -ffunction-sections -fdata-sections -fcoverage-compilation-dir=/build/source/build-llvm/tools/clang/stage2-bins -resource-dir /usr/lib/llvm-16/lib/clang/16.0.0 -D MLIR_CUDA_CONVERSIONS_ENABLED=1 -D MLIR_ROCM_CONVERSIONS_ENABLED=1 -D _DEBUG -D _GNU_SOURCE -D __STDC_CONSTANT_MACROS -D __STDC_FORMAT_MACROS -D __STDC_LIMIT_MACROS -I tools/mlir/lib/Dialect/SPIRV/IR -I /build/source/mlir/lib/Dialect/SPIRV/IR -I include -I /build/source/llvm/include -I /build/source/mlir/include -I tools/mlir/include -D _FORTIFY_SOURCE=2 -D NDEBUG -U NDEBUG -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/10/../../../../include/c++/10 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/10/../../../../include/x86_64-linux-gnu/c++/10 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/10/../../../../include/c++/10/backward -internal-isystem /usr/lib/llvm-16/lib/clang/16.0.0/include -internal-isystem /usr/local/include -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/10/../../../../x86_64-linux-gnu/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -fmacro-prefix-map=/build/source/build-llvm/tools/clang/stage2-bins=build-llvm/tools/clang/stage2-bins -fmacro-prefix-map=/build/source/= -fcoverage-prefix-map=/build/source/build-llvm/tools/clang/stage2-bins=build-llvm/tools/clang/stage2-bins -fcoverage-prefix-map=/build/source/= -source-date-epoch 1668078801 -O2 -Wno-unused-command-line-argument -Wno-unused-parameter -Wwrite-strings -Wno-missing-field-initializers -Wno-long-long -Wno-maybe-uninitialized -Wno-class-memaccess -Wno-redundant-move -Wno-pessimizing-move -Wno-noexcept-type -Wno-comment -Wno-misleading-indentation -std=c++17 -fdeprecated-macro -fdebug-compilation-dir=/build/source/build-llvm/tools/clang/stage2-bins -fdebug-prefix-map=/build/source/build-llvm/tools/clang/stage2-bins=build-llvm/tools/clang/stage2-bins -fdebug-prefix-map=/build/source/= -ferror-limit 19 -fvisibility-inlines-hidden -stack-protector 2 -fgnuc-version=4.2.1 -fcolor-diagnostics -vectorize-loops -vectorize-slp -analyzer-output=html -analyzer-config stable-report-filename=true -faddrsig -D__GCC_HAVE_DWARF2_CFI_ASM=1 -o /tmp/scan-build-2022-11-10-135928-647445-1 -x c++ /build/source/mlir/lib/Dialect/SPIRV/IR/SPIRVAttributes.cpp
1/*===- TableGen'erated file -------------------------------------*- C++ -*-===*\
2|* *|
3|* AttrDef Definitions *|
4|* *|
5|* Automatically generated file, do not edit! *|
6|* *|
7\*===----------------------------------------------------------------------===*/
8
9#ifdef GET_ATTRDEF_LIST
10#undef GET_ATTRDEF_LIST
11
12::mlir::spirv::AddressingModelAttr,
13::mlir::spirv::ImageArrayedInfoAttr,
14::mlir::spirv::BuiltInAttr,
15::mlir::spirv::CapabilityAttr,
16::mlir::spirv::CooperativeMatrixPropertiesNVAttr,
17::mlir::spirv::DecorationAttr,
18::mlir::spirv::ImageDepthInfoAttr,
19::mlir::spirv::DeviceTypeAttr,
20::mlir::spirv::DimAttr,
21::mlir::spirv::EntryPointABIAttr,
22::mlir::spirv::ExecutionModeAttr,
23::mlir::spirv::ExecutionModelAttr,
24::mlir::spirv::ExtensionAttr,
25::mlir::spirv::FunctionControlAttr,
26::mlir::spirv::GroupOperationAttr,
27::mlir::spirv::ImageFormatAttr,
28::mlir::spirv::ImageOperandsAttr,
29::mlir::spirv::JointMatrixPropertiesINTELAttr,
30::mlir::spirv::LinkageTypeAttr,
31::mlir::spirv::LoopControlAttr,
32::mlir::spirv::MatrixLayoutAttr,
33::mlir::spirv::MemoryAccessAttr,
34::mlir::spirv::MemoryModelAttr,
35::mlir::spirv::MemorySemanticsAttr,
36::mlir::spirv::OpcodeAttr,
37::mlir::spirv::ResourceLimitsAttr,
38::mlir::spirv::ImageSamplerUseInfoAttr,
39::mlir::spirv::ImageSamplingInfoAttr,
40::mlir::spirv::ScopeAttr,
41::mlir::spirv::SelectionControlAttr,
42::mlir::spirv::StorageClassAttr,
43::mlir::spirv::VendorAttr,
44::mlir::spirv::VersionAttr
45
46#endif // GET_ATTRDEF_LIST
47
48#ifdef GET_ATTRDEF_CLASSES
49#undef GET_ATTRDEF_CLASSES
50
51static ::mlir::OptionalParseResult generatedAttributeParser(::mlir::AsmParser &parser, ::llvm::StringRef *mnemonic, ::mlir::Type type, ::mlir::Attribute &value) {
52 return ::mlir::AsmParser::KeywordSwitch<::mlir::OptionalParseResult>(parser)
53 .Case(::mlir::spirv::AddressingModelAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
54 value = ::mlir::spirv::AddressingModelAttr::parse(parser, type);
55 return ::mlir::success(!!value);
56 })
57 .Case(::mlir::spirv::ImageArrayedInfoAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
58 value = ::mlir::spirv::ImageArrayedInfoAttr::parse(parser, type);
59 return ::mlir::success(!!value);
60 })
61 .Case(::mlir::spirv::BuiltInAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
62 value = ::mlir::spirv::BuiltInAttr::parse(parser, type);
63 return ::mlir::success(!!value);
64 })
65 .Case(::mlir::spirv::CapabilityAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
66 value = ::mlir::spirv::CapabilityAttr::parse(parser, type);
67 return ::mlir::success(!!value);
68 })
69 .Case(::mlir::spirv::CooperativeMatrixPropertiesNVAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
70 value = ::mlir::spirv::CooperativeMatrixPropertiesNVAttr::parse(parser, type);
71 return ::mlir::success(!!value);
72 })
73 .Case(::mlir::spirv::DecorationAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
74 value = ::mlir::spirv::DecorationAttr::parse(parser, type);
75 return ::mlir::success(!!value);
76 })
77 .Case(::mlir::spirv::ImageDepthInfoAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
78 value = ::mlir::spirv::ImageDepthInfoAttr::parse(parser, type);
79 return ::mlir::success(!!value);
80 })
81 .Case(::mlir::spirv::DeviceTypeAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
82 value = ::mlir::spirv::DeviceTypeAttr::parse(parser, type);
83 return ::mlir::success(!!value);
84 })
85 .Case(::mlir::spirv::DimAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
86 value = ::mlir::spirv::DimAttr::parse(parser, type);
87 return ::mlir::success(!!value);
88 })
89 .Case(::mlir::spirv::EntryPointABIAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
90 value = ::mlir::spirv::EntryPointABIAttr::parse(parser, type);
91 return ::mlir::success(!!value);
92 })
93 .Case(::mlir::spirv::ExecutionModeAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
94 value = ::mlir::spirv::ExecutionModeAttr::parse(parser, type);
95 return ::mlir::success(!!value);
96 })
97 .Case(::mlir::spirv::ExecutionModelAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
98 value = ::mlir::spirv::ExecutionModelAttr::parse(parser, type);
99 return ::mlir::success(!!value);
100 })
101 .Case(::mlir::spirv::ExtensionAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
102 value = ::mlir::spirv::ExtensionAttr::parse(parser, type);
103 return ::mlir::success(!!value);
104 })
105 .Case(::mlir::spirv::FunctionControlAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
106 value = ::mlir::spirv::FunctionControlAttr::parse(parser, type);
107 return ::mlir::success(!!value);
108 })
109 .Case(::mlir::spirv::GroupOperationAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
110 value = ::mlir::spirv::GroupOperationAttr::parse(parser, type);
111 return ::mlir::success(!!value);
112 })
113 .Case(::mlir::spirv::ImageFormatAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
114 value = ::mlir::spirv::ImageFormatAttr::parse(parser, type);
115 return ::mlir::success(!!value);
116 })
117 .Case(::mlir::spirv::ImageOperandsAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
118 value = ::mlir::spirv::ImageOperandsAttr::parse(parser, type);
119 return ::mlir::success(!!value);
120 })
121 .Case(::mlir::spirv::JointMatrixPropertiesINTELAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
122 value = ::mlir::spirv::JointMatrixPropertiesINTELAttr::parse(parser, type);
123 return ::mlir::success(!!value);
124 })
125 .Case(::mlir::spirv::LinkageTypeAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
126 value = ::mlir::spirv::LinkageTypeAttr::parse(parser, type);
127 return ::mlir::success(!!value);
128 })
129 .Case(::mlir::spirv::LoopControlAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
130 value = ::mlir::spirv::LoopControlAttr::parse(parser, type);
131 return ::mlir::success(!!value);
132 })
133 .Case(::mlir::spirv::MatrixLayoutAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
134 value = ::mlir::spirv::MatrixLayoutAttr::parse(parser, type);
135 return ::mlir::success(!!value);
136 })
137 .Case(::mlir::spirv::MemoryAccessAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
138 value = ::mlir::spirv::MemoryAccessAttr::parse(parser, type);
139 return ::mlir::success(!!value);
140 })
141 .Case(::mlir::spirv::MemoryModelAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
142 value = ::mlir::spirv::MemoryModelAttr::parse(parser, type);
143 return ::mlir::success(!!value);
144 })
145 .Case(::mlir::spirv::MemorySemanticsAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
146 value = ::mlir::spirv::MemorySemanticsAttr::parse(parser, type);
147 return ::mlir::success(!!value);
148 })
149 .Case(::mlir::spirv::OpcodeAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
150 value = ::mlir::spirv::OpcodeAttr::parse(parser, type);
151 return ::mlir::success(!!value);
152 })
153 .Case(::mlir::spirv::ResourceLimitsAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
154 value = ::mlir::spirv::ResourceLimitsAttr::parse(parser, type);
155 return ::mlir::success(!!value);
156 })
157 .Case(::mlir::spirv::ImageSamplerUseInfoAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
158 value = ::mlir::spirv::ImageSamplerUseInfoAttr::parse(parser, type);
159 return ::mlir::success(!!value);
160 })
161 .Case(::mlir::spirv::ImageSamplingInfoAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
162 value = ::mlir::spirv::ImageSamplingInfoAttr::parse(parser, type);
163 return ::mlir::success(!!value);
164 })
165 .Case(::mlir::spirv::ScopeAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
166 value = ::mlir::spirv::ScopeAttr::parse(parser, type);
167 return ::mlir::success(!!value);
168 })
169 .Case(::mlir::spirv::SelectionControlAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
170 value = ::mlir::spirv::SelectionControlAttr::parse(parser, type);
171 return ::mlir::success(!!value);
172 })
173 .Case(::mlir::spirv::StorageClassAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
174 value = ::mlir::spirv::StorageClassAttr::parse(parser, type);
175 return ::mlir::success(!!value);
176 })
177 .Case(::mlir::spirv::VendorAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
178 value = ::mlir::spirv::VendorAttr::parse(parser, type);
179 return ::mlir::success(!!value);
180 })
181 .Case(::mlir::spirv::VersionAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
182 value = ::mlir::spirv::VersionAttr::parse(parser, type);
183 return ::mlir::success(!!value);
184 })
185 .Default([&](llvm::StringRef keyword, llvm::SMLoc) {
186 *mnemonic = keyword;
187 return llvm::None;
188 });
189}
190
191static ::mlir::LogicalResult generatedAttributePrinter(::mlir::Attribute def, ::mlir::AsmPrinter &printer) {
192 return ::llvm::TypeSwitch<::mlir::Attribute, ::mlir::LogicalResult>(def) .Case<::mlir::spirv::AddressingModelAttr>([&](auto t) {
193 printer << ::mlir::spirv::AddressingModelAttr::getMnemonic();
194t.print(printer);
195 return ::mlir::success();
196 })
197 .Case<::mlir::spirv::ImageArrayedInfoAttr>([&](auto t) {
198 printer << ::mlir::spirv::ImageArrayedInfoAttr::getMnemonic();
199t.print(printer);
200 return ::mlir::success();
201 })
202 .Case<::mlir::spirv::BuiltInAttr>([&](auto t) {
203 printer << ::mlir::spirv::BuiltInAttr::getMnemonic();
204t.print(printer);
205 return ::mlir::success();
206 })
207 .Case<::mlir::spirv::CapabilityAttr>([&](auto t) {
208 printer << ::mlir::spirv::CapabilityAttr::getMnemonic();
209t.print(printer);
210 return ::mlir::success();
211 })
212 .Case<::mlir::spirv::CooperativeMatrixPropertiesNVAttr>([&](auto t) {
213 printer << ::mlir::spirv::CooperativeMatrixPropertiesNVAttr::getMnemonic();
214t.print(printer);
215 return ::mlir::success();
216 })
217 .Case<::mlir::spirv::DecorationAttr>([&](auto t) {
218 printer << ::mlir::spirv::DecorationAttr::getMnemonic();
219t.print(printer);
220 return ::mlir::success();
221 })
222 .Case<::mlir::spirv::ImageDepthInfoAttr>([&](auto t) {
223 printer << ::mlir::spirv::ImageDepthInfoAttr::getMnemonic();
224t.print(printer);
225 return ::mlir::success();
226 })
227 .Case<::mlir::spirv::DeviceTypeAttr>([&](auto t) {
228 printer << ::mlir::spirv::DeviceTypeAttr::getMnemonic();
229t.print(printer);
230 return ::mlir::success();
231 })
232 .Case<::mlir::spirv::DimAttr>([&](auto t) {
233 printer << ::mlir::spirv::DimAttr::getMnemonic();
234t.print(printer);
235 return ::mlir::success();
236 })
237 .Case<::mlir::spirv::EntryPointABIAttr>([&](auto t) {
238 printer << ::mlir::spirv::EntryPointABIAttr::getMnemonic();
239t.print(printer);
240 return ::mlir::success();
241 })
242 .Case<::mlir::spirv::ExecutionModeAttr>([&](auto t) {
243 printer << ::mlir::spirv::ExecutionModeAttr::getMnemonic();
244t.print(printer);
245 return ::mlir::success();
246 })
247 .Case<::mlir::spirv::ExecutionModelAttr>([&](auto t) {
248 printer << ::mlir::spirv::ExecutionModelAttr::getMnemonic();
249t.print(printer);
250 return ::mlir::success();
251 })
252 .Case<::mlir::spirv::ExtensionAttr>([&](auto t) {
253 printer << ::mlir::spirv::ExtensionAttr::getMnemonic();
254t.print(printer);
255 return ::mlir::success();
256 })
257 .Case<::mlir::spirv::FunctionControlAttr>([&](auto t) {
258 printer << ::mlir::spirv::FunctionControlAttr::getMnemonic();
259t.print(printer);
260 return ::mlir::success();
261 })
262 .Case<::mlir::spirv::GroupOperationAttr>([&](auto t) {
263 printer << ::mlir::spirv::GroupOperationAttr::getMnemonic();
264t.print(printer);
265 return ::mlir::success();
266 })
267 .Case<::mlir::spirv::ImageFormatAttr>([&](auto t) {
268 printer << ::mlir::spirv::ImageFormatAttr::getMnemonic();
269t.print(printer);
270 return ::mlir::success();
271 })
272 .Case<::mlir::spirv::ImageOperandsAttr>([&](auto t) {
273 printer << ::mlir::spirv::ImageOperandsAttr::getMnemonic();
274t.print(printer);
275 return ::mlir::success();
276 })
277 .Case<::mlir::spirv::JointMatrixPropertiesINTELAttr>([&](auto t) {
278 printer << ::mlir::spirv::JointMatrixPropertiesINTELAttr::getMnemonic();
279t.print(printer);
280 return ::mlir::success();
281 })
282 .Case<::mlir::spirv::LinkageTypeAttr>([&](auto t) {
283 printer << ::mlir::spirv::LinkageTypeAttr::getMnemonic();
284t.print(printer);
285 return ::mlir::success();
286 })
287 .Case<::mlir::spirv::LoopControlAttr>([&](auto t) {
288 printer << ::mlir::spirv::LoopControlAttr::getMnemonic();
289t.print(printer);
290 return ::mlir::success();
291 })
292 .Case<::mlir::spirv::MatrixLayoutAttr>([&](auto t) {
293 printer << ::mlir::spirv::MatrixLayoutAttr::getMnemonic();
294t.print(printer);
295 return ::mlir::success();
296 })
297 .Case<::mlir::spirv::MemoryAccessAttr>([&](auto t) {
298 printer << ::mlir::spirv::MemoryAccessAttr::getMnemonic();
299t.print(printer);
300 return ::mlir::success();
301 })
302 .Case<::mlir::spirv::MemoryModelAttr>([&](auto t) {
303 printer << ::mlir::spirv::MemoryModelAttr::getMnemonic();
304t.print(printer);
305 return ::mlir::success();
306 })
307 .Case<::mlir::spirv::MemorySemanticsAttr>([&](auto t) {
308 printer << ::mlir::spirv::MemorySemanticsAttr::getMnemonic();
309t.print(printer);
310 return ::mlir::success();
311 })
312 .Case<::mlir::spirv::OpcodeAttr>([&](auto t) {
313 printer << ::mlir::spirv::OpcodeAttr::getMnemonic();
314t.print(printer);
315 return ::mlir::success();
316 })
317 .Case<::mlir::spirv::ResourceLimitsAttr>([&](auto t) {
318 printer << ::mlir::spirv::ResourceLimitsAttr::getMnemonic();
319t.print(printer);
320 return ::mlir::success();
321 })
322 .Case<::mlir::spirv::ImageSamplerUseInfoAttr>([&](auto t) {
323 printer << ::mlir::spirv::ImageSamplerUseInfoAttr::getMnemonic();
324t.print(printer);
325 return ::mlir::success();
326 })
327 .Case<::mlir::spirv::ImageSamplingInfoAttr>([&](auto t) {
328 printer << ::mlir::spirv::ImageSamplingInfoAttr::getMnemonic();
329t.print(printer);
330 return ::mlir::success();
331 })
332 .Case<::mlir::spirv::ScopeAttr>([&](auto t) {
333 printer << ::mlir::spirv::ScopeAttr::getMnemonic();
334t.print(printer);
335 return ::mlir::success();
336 })
337 .Case<::mlir::spirv::SelectionControlAttr>([&](auto t) {
338 printer << ::mlir::spirv::SelectionControlAttr::getMnemonic();
339t.print(printer);
340 return ::mlir::success();
341 })
342 .Case<::mlir::spirv::StorageClassAttr>([&](auto t) {
343 printer << ::mlir::spirv::StorageClassAttr::getMnemonic();
344t.print(printer);
345 return ::mlir::success();
346 })
347 .Case<::mlir::spirv::VendorAttr>([&](auto t) {
348 printer << ::mlir::spirv::VendorAttr::getMnemonic();
349t.print(printer);
350 return ::mlir::success();
351 })
352 .Case<::mlir::spirv::VersionAttr>([&](auto t) {
353 printer << ::mlir::spirv::VersionAttr::getMnemonic();
354t.print(printer);
355 return ::mlir::success();
356 })
357 .Default([](auto) { return ::mlir::failure(); });
358}
359
360namespace mlir {
361namespace spirv {
362namespace detail {
363struct AddressingModelAttrStorage : public ::mlir::AttributeStorage {
364 using KeyTy = std::tuple<::mlir::spirv::AddressingModel>;
365 AddressingModelAttrStorage(::mlir::spirv::AddressingModel value) : value(value) {}
366
367 KeyTy getAsKey() const {
368 return KeyTy(value);
369 }
370
371 bool operator==(const KeyTy &tblgenKey) const {
372 return (value == std::get<0>(tblgenKey));
373 }
374
375 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
376 return ::llvm::hash_combine(std::get<0>(tblgenKey));
377 }
378
379 static AddressingModelAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
380 auto value = std::get<0>(tblgenKey);
381 return new (allocator.allocate<AddressingModelAttrStorage>()) AddressingModelAttrStorage(value);
382 }
383
384 ::mlir::spirv::AddressingModel value;
385};
386} // namespace detail
387AddressingModelAttr AddressingModelAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::AddressingModel value) {
388 return Base::get(context, value);
389}
390
391::mlir::Attribute AddressingModelAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
392 ::mlir::Builder odsBuilder(odsParser.getContext());
393 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
394 (void) odsLoc;
395 ::mlir::FailureOr<::mlir::spirv::AddressingModel> _result_value;
396 // Parse literal '<'
397 if (odsParser.parseLess()) return {};
398
399 // Parse variable 'value'
400 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::AddressingModel> {
401 auto loc = odsParser.getCurrentLocation();
402 ::llvm::StringRef enumKeyword;
403 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
404 return ::mlir::failure();
405 auto maybeEnum = ::mlir::spirv::symbolizeAddressingModel(enumKeyword);
406 if (maybeEnum)
407 return *maybeEnum;
408 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::AddressingModel" << " to be one of: " << "Logical" << ", " << "Physical32" << ", " << "Physical64" << ", " << "PhysicalStorageBuffer64")};
409 }();
410 if (::mlir::failed(_result_value)) {
411 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_AddressingModelAttr parameter 'value' which is to be a `::mlir::spirv::AddressingModel`");
412 return {};
413 }
414 // Parse literal '>'
415 if (odsParser.parseGreater()) return {};
416 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 416, __extension__ __PRETTY_FUNCTION__));
417 return AddressingModelAttr::get(odsParser.getContext(),
418 ::mlir::spirv::AddressingModel((*_result_value)));
419}
420
421void AddressingModelAttr::print(::mlir::AsmPrinter &odsPrinter) const {
422 ::mlir::Builder odsBuilder(getContext());
423 odsPrinter << "<";
424 odsPrinter << stringifyAddressingModel(getValue());
425 odsPrinter << ">";
426}
427
428::mlir::spirv::AddressingModel AddressingModelAttr::getValue() const {
429 return getImpl()->value;
430}
431
432} // namespace spirv
433} // namespace mlir
434MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::AddressingModelAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::AddressingModelAttr>::id = {}; } }
435namespace mlir {
436namespace spirv {
437namespace detail {
438struct ImageArrayedInfoAttrStorage : public ::mlir::AttributeStorage {
439 using KeyTy = std::tuple<::mlir::spirv::ImageArrayedInfo>;
440 ImageArrayedInfoAttrStorage(::mlir::spirv::ImageArrayedInfo value) : value(value) {}
441
442 KeyTy getAsKey() const {
443 return KeyTy(value);
444 }
445
446 bool operator==(const KeyTy &tblgenKey) const {
447 return (value == std::get<0>(tblgenKey));
448 }
449
450 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
451 return ::llvm::hash_combine(std::get<0>(tblgenKey));
452 }
453
454 static ImageArrayedInfoAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
455 auto value = std::get<0>(tblgenKey);
456 return new (allocator.allocate<ImageArrayedInfoAttrStorage>()) ImageArrayedInfoAttrStorage(value);
457 }
458
459 ::mlir::spirv::ImageArrayedInfo value;
460};
461} // namespace detail
462ImageArrayedInfoAttr ImageArrayedInfoAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ImageArrayedInfo value) {
463 return Base::get(context, value);
464}
465
466::mlir::Attribute ImageArrayedInfoAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
467 ::mlir::Builder odsBuilder(odsParser.getContext());
468 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
469 (void) odsLoc;
470 ::mlir::FailureOr<::mlir::spirv::ImageArrayedInfo> _result_value;
471 // Parse literal '<'
472 if (odsParser.parseLess()) return {};
473
474 // Parse variable 'value'
475 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ImageArrayedInfo> {
476 auto loc = odsParser.getCurrentLocation();
477 ::llvm::StringRef enumKeyword;
478 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
479 return ::mlir::failure();
480 auto maybeEnum = ::mlir::spirv::symbolizeImageArrayedInfo(enumKeyword);
481 if (maybeEnum)
482 return *maybeEnum;
483 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ImageArrayedInfo" << " to be one of: " << "NonArrayed" << ", " << "Arrayed")};
484 }();
485 if (::mlir::failed(_result_value)) {
486 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ArrayedAttr parameter 'value' which is to be a `::mlir::spirv::ImageArrayedInfo`");
487 return {};
488 }
489 // Parse literal '>'
490 if (odsParser.parseGreater()) return {};
491 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 491, __extension__ __PRETTY_FUNCTION__));
492 return ImageArrayedInfoAttr::get(odsParser.getContext(),
493 ::mlir::spirv::ImageArrayedInfo((*_result_value)));
494}
495
496void ImageArrayedInfoAttr::print(::mlir::AsmPrinter &odsPrinter) const {
497 ::mlir::Builder odsBuilder(getContext());
498 odsPrinter << "<";
499 odsPrinter << stringifyImageArrayedInfo(getValue());
500 odsPrinter << ">";
501}
502
503::mlir::spirv::ImageArrayedInfo ImageArrayedInfoAttr::getValue() const {
504 return getImpl()->value;
505}
506
507} // namespace spirv
508} // namespace mlir
509MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ImageArrayedInfoAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ImageArrayedInfoAttr>::id = {}; } }
510namespace mlir {
511namespace spirv {
512namespace detail {
513struct BuiltInAttrStorage : public ::mlir::AttributeStorage {
514 using KeyTy = std::tuple<::mlir::spirv::BuiltIn>;
515 BuiltInAttrStorage(::mlir::spirv::BuiltIn value) : value(value) {}
516
517 KeyTy getAsKey() const {
518 return KeyTy(value);
519 }
520
521 bool operator==(const KeyTy &tblgenKey) const {
522 return (value == std::get<0>(tblgenKey));
523 }
524
525 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
526 return ::llvm::hash_combine(std::get<0>(tblgenKey));
527 }
528
529 static BuiltInAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
530 auto value = std::get<0>(tblgenKey);
531 return new (allocator.allocate<BuiltInAttrStorage>()) BuiltInAttrStorage(value);
532 }
533
534 ::mlir::spirv::BuiltIn value;
535};
536} // namespace detail
537BuiltInAttr BuiltInAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::BuiltIn value) {
538 return Base::get(context, value);
539}
540
541::mlir::Attribute BuiltInAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
542 ::mlir::Builder odsBuilder(odsParser.getContext());
543 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
544 (void) odsLoc;
545 ::mlir::FailureOr<::mlir::spirv::BuiltIn> _result_value;
546 // Parse literal '<'
547 if (odsParser.parseLess()) return {};
548
549 // Parse variable 'value'
550 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::BuiltIn> {
551 auto loc = odsParser.getCurrentLocation();
552 ::llvm::StringRef enumKeyword;
553 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
554 return ::mlir::failure();
555 auto maybeEnum = ::mlir::spirv::symbolizeBuiltIn(enumKeyword);
556 if (maybeEnum)
557 return *maybeEnum;
558 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::BuiltIn" << " to be one of: " << "Position" << ", " << "PointSize" << ", " << "ClipDistance" << ", " << "CullDistance" << ", " << "VertexId" << ", " << "InstanceId" << ", " << "PrimitiveId" << ", " << "InvocationId" << ", " << "Layer" << ", " << "ViewportIndex" << ", " << "TessLevelOuter" << ", " << "TessLevelInner" << ", " << "TessCoord" << ", " << "PatchVertices" << ", " << "FragCoord" << ", " << "PointCoord" << ", " << "FrontFacing" << ", " << "SampleId" << ", " << "SamplePosition" << ", " << "SampleMask" << ", " << "FragDepth" << ", " << "HelperInvocation" << ", " << "NumWorkgroups" << ", " << "WorkgroupSize" << ", " << "WorkgroupId" << ", " << "LocalInvocationId" << ", " << "GlobalInvocationId" << ", " << "LocalInvocationIndex" << ", " << "WorkDim" << ", " << "GlobalSize" << ", " << "EnqueuedWorkgroupSize" << ", " << "GlobalOffset" << ", " << "GlobalLinearId" << ", " << "SubgroupSize" << ", " << "SubgroupMaxSize" << ", " << "NumSubgroups" << ", " << "NumEnqueuedSubgroups" << ", " << "SubgroupId" << ", " << "SubgroupLocalInvocationId" << ", " << "VertexIndex" << ", " << "InstanceIndex" << ", " << "SubgroupEqMask" << ", " << "SubgroupGeMask" << ", " << "SubgroupGtMask" << ", " << "SubgroupLeMask" << ", " << "SubgroupLtMask" << ", " << "BaseVertex" << ", " << "BaseInstance" << ", " << "DrawIndex" << ", " << "PrimitiveShadingRateKHR" << ", " << "DeviceIndex" << ", " << "ViewIndex" << ", " << "ShadingRateKHR" << ", " << "BaryCoordNoPerspAMD" << ", " << "BaryCoordNoPerspCentroidAMD" << ", " << "BaryCoordNoPerspSampleAMD" << ", " << "BaryCoordSmoothAMD" << ", " << "BaryCoordSmoothCentroidAMD" << ", " << "BaryCoordSmoothSampleAMD" << ", " << "BaryCoordPullModelAMD" << ", " << "FragStencilRefEXT" << ", " << "ViewportMaskNV" << ", " << "SecondaryPositionNV" << ", " << "SecondaryViewportMaskNV" << ", " << "PositionPerViewNV" << ", " << "ViewportMaskPerViewNV" << ", " << "FullyCoveredEXT" << ", " << "TaskCountNV" << ", " << "PrimitiveCountNV" << ", " << "PrimitiveIndicesNV" << ", " << "ClipDistancePerViewNV" << ", " << "CullDistancePerViewNV" << ", " << "LayerPerViewNV" << ", " << "MeshViewCountNV" << ", " << "MeshViewIndicesNV" << ", " << "BaryCoordKHR" << ", " << "BaryCoordNoPerspKHR" << ", " << "FragSizeEXT" << ", " << "FragInvocationCountEXT" << ", " << "LaunchIdKHR" << ", " << "LaunchSizeKHR" << ", " << "WorldRayOriginKHR" << ", " << "WorldRayDirectionKHR" << ", " << "ObjectRayOriginKHR" << ", " << "ObjectRayDirectionKHR" << ", " << "RayTminKHR" << ", " << "RayTmaxKHR" << ", " << "InstanceCustomIndexKHR" << ", " << "ObjectToWorldKHR" << ", " << "WorldToObjectKHR" << ", " << "HitTNV" << ", " << "HitKindKHR" << ", " << "CurrentRayTimeNV" << ", " << "IncomingRayFlagsKHR" << ", " << "RayGeometryIndexKHR" << ", " << "WarpsPerSMNV" << ", " << "SMCountNV" << ", " << "WarpIDNV" << ", " << "SMIDNV" << ", " << "CullMaskKHR")};
559 }();
560 if (::mlir::failed(_result_value)) {
561 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_BuiltInAttr parameter 'value' which is to be a `::mlir::spirv::BuiltIn`");
562 return {};
563 }
564 // Parse literal '>'
565 if (odsParser.parseGreater()) return {};
566 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 566, __extension__ __PRETTY_FUNCTION__));
567 return BuiltInAttr::get(odsParser.getContext(),
568 ::mlir::spirv::BuiltIn((*_result_value)));
569}
570
571void BuiltInAttr::print(::mlir::AsmPrinter &odsPrinter) const {
572 ::mlir::Builder odsBuilder(getContext());
573 odsPrinter << "<";
574 odsPrinter << stringifyBuiltIn(getValue());
575 odsPrinter << ">";
576}
577
578::mlir::spirv::BuiltIn BuiltInAttr::getValue() const {
579 return getImpl()->value;
580}
581
582} // namespace spirv
583} // namespace mlir
584MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::BuiltInAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::BuiltInAttr>::id = {}; } }
585namespace mlir {
586namespace spirv {
587namespace detail {
588struct CapabilityAttrStorage : public ::mlir::AttributeStorage {
589 using KeyTy = std::tuple<::mlir::spirv::Capability>;
590 CapabilityAttrStorage(::mlir::spirv::Capability value) : value(value) {}
591
592 KeyTy getAsKey() const {
593 return KeyTy(value);
594 }
595
596 bool operator==(const KeyTy &tblgenKey) const {
597 return (value == std::get<0>(tblgenKey));
598 }
599
600 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
601 return ::llvm::hash_combine(std::get<0>(tblgenKey));
602 }
603
604 static CapabilityAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
605 auto value = std::get<0>(tblgenKey);
606 return new (allocator.allocate<CapabilityAttrStorage>()) CapabilityAttrStorage(value);
607 }
608
609 ::mlir::spirv::Capability value;
610};
611} // namespace detail
612CapabilityAttr CapabilityAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Capability value) {
613 return Base::get(context, value);
614}
615
616::mlir::Attribute CapabilityAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
617 ::mlir::Builder odsBuilder(odsParser.getContext());
618 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
619 (void) odsLoc;
620 ::mlir::FailureOr<::mlir::spirv::Capability> _result_value;
621 // Parse literal '<'
622 if (odsParser.parseLess()) return {};
623
624 // Parse variable 'value'
625 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Capability> {
626 auto loc = odsParser.getCurrentLocation();
627 ::llvm::StringRef enumKeyword;
628 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
629 return ::mlir::failure();
630 auto maybeEnum = ::mlir::spirv::symbolizeCapability(enumKeyword);
631 if (maybeEnum)
632 return *maybeEnum;
633 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::Capability" << " to be one of: " << "Matrix" << ", " << "Addresses" << ", " << "Linkage" << ", " << "Kernel" << ", " << "Float16" << ", " << "Float64" << ", " << "Int64" << ", " << "Groups" << ", " << "Int16" << ", " << "Int8" << ", " << "Sampled1D" << ", " << "SampledBuffer" << ", " << "GroupNonUniform" << ", " << "ShaderLayer" << ", " << "ShaderViewportIndex" << ", " << "UniformDecoration" << ", " << "SubgroupBallotKHR" << ", " << "SubgroupVoteKHR" << ", " << "StorageBuffer16BitAccess" << ", " << "StoragePushConstant16" << ", " << "StorageInputOutput16" << ", " << "DeviceGroup" << ", " << "AtomicStorageOps" << ", " << "SampleMaskPostDepthCoverage" << ", " << "StorageBuffer8BitAccess" << ", " << "StoragePushConstant8" << ", " << "DenormPreserve" << ", " << "DenormFlushToZero" << ", " << "SignedZeroInfNanPreserve" << ", " << "RoundingModeRTE" << ", " << "RoundingModeRTZ" << ", " << "ImageFootprintNV" << ", " << "FragmentBarycentricKHR" << ", " << "ComputeDerivativeGroupQuadsNV" << ", " << "GroupNonUniformPartitionedNV" << ", " << "VulkanMemoryModel" << ", " << "VulkanMemoryModelDeviceScope" << ", " << "ComputeDerivativeGroupLinearNV" << ", " << "BindlessTextureNV" << ", " << "SubgroupShuffleINTEL" << ", " << "SubgroupBufferBlockIOINTEL" << ", " << "SubgroupImageBlockIOINTEL" << ", " << "SubgroupImageMediaBlockIOINTEL" << ", " << "RoundToInfinityINTEL" << ", " << "FloatingPointModeINTEL" << ", " << "FunctionPointersINTEL" << ", " << "IndirectReferencesINTEL" << ", " << "AsmINTEL" << ", " << "AtomicFloat32MinMaxEXT" << ", " << "AtomicFloat64MinMaxEXT" << ", " << "AtomicFloat16MinMaxEXT" << ", " << "VectorAnyINTEL" << ", " << "ExpectAssumeKHR" << ", " << "SubgroupAvcMotionEstimationINTEL" << ", " << "SubgroupAvcMotionEstimationIntraINTEL" << ", " << "SubgroupAvcMotionEstimationChromaINTEL" << ", " << "VariableLengthArrayINTEL" << ", " << "FunctionFloatControlINTEL" << ", " << "FPGAMemoryAttributesINTEL" << ", " << "ArbitraryPrecisionIntegersINTEL" << ", " << "ArbitraryPrecisionFloatingPointINTEL" << ", " << "UnstructuredLoopControlsINTEL" << ", " << "FPGALoopControlsINTEL" << ", " << "KernelAttributesINTEL" << ", " << "FPGAKernelAttributesINTEL" << ", " << "FPGAMemoryAccessesINTEL" << ", " << "FPGAClusterAttributesINTEL" << ", " << "LoopFuseINTEL" << ", " << "MemoryAccessAliasingINTEL" << ", " << "FPGABufferLocationINTEL" << ", " << "ArbitraryPrecisionFixedPointINTEL" << ", " << "USMStorageClassesINTEL" << ", " << "IOPipesINTEL" << ", " << "BlockingPipesINTEL" << ", " << "FPGARegINTEL" << ", " << "DotProductInputAll" << ", " << "DotProductInput4x8BitPacked" << ", " << "DotProduct" << ", " << "RayCullMaskKHR" << ", " << "BitInstructions" << ", " << "AtomicFloat32AddEXT" << ", " << "AtomicFloat64AddEXT" << ", " << "LongConstantCompositeINTEL" << ", " << "OptNoneINTEL" << ", " << "AtomicFloat16AddEXT" << ", " << "DebugInfoModuleINTEL" << ", " << "SplitBarrierINTEL" << ", " << "GroupUniformArithmeticKHR" << ", " << "Shader" << ", " << "Vector16" << ", " << "Float16Buffer" << ", " << "Int64Atomics" << ", " << "ImageBasic" << ", " << "Pipes" << ", " << "DeviceEnqueue" << ", " << "LiteralSampler" << ", " << "GenericPointer" << ", " << "Image1D" << ", " << "ImageBuffer" << ", " << "NamedBarrier" << ", " << "GroupNonUniformVote" << ", " << "GroupNonUniformArithmetic" << ", " << "GroupNonUniformBallot" << ", " << "GroupNonUniformShuffle" << ", " << "GroupNonUniformShuffleRelative" << ", " << "GroupNonUniformClustered" << ", " << "GroupNonUniformQuad" << ", " << "StorageUniform16" << ", " << "UniformAndStorageBuffer8BitAccess" << ", " << "UniformTexelBufferArrayDynamicIndexing" << ", " << "VectorComputeINTEL" << ", " << "FPFastMathModeINTEL" << ", " << "DotProductInput4x8Bit" << ", " << "GroupNonUniformRotateKHR" << ", " << "Geometry" << ", " << "Tessellation" << ", " << "ImageReadWrite" << ", " << "ImageMipmap" << ", " << "AtomicStorage" << ", " << "ImageGatherExtended" << ", " << "StorageImageMultisample" << ", " << "UniformBufferArrayDynamicIndexing" << ", " << "SampledImageArrayDynamicIndexing" << ", " << "StorageBufferArrayDynamicIndexing" << ", " << "StorageImageArrayDynamicIndexing" << ", " << "ClipDistance" << ", " << "CullDistance" << ", " << "SampleRateShading" << ", " << "SampledRect" << ", " << "InputAttachment" << ", " << "SparseResidency" << ", " << "MinLod" << ", " << "SampledCubeArray" << ", " << "ImageMSArray" << ", " << "StorageImageExtendedFormats" << ", " << "ImageQuery" << ", " << "DerivativeControl" << ", " << "InterpolationFunction" << ", " << "TransformFeedback" << ", " << "StorageImageReadWithoutFormat" << ", " << "StorageImageWriteWithoutFormat" << ", " << "SubgroupDispatch" << ", " << "PipeStorage" << ", " << "FragmentShadingRateKHR" << ", " << "DrawParameters" << ", " << "WorkgroupMemoryExplicitLayoutKHR" << ", " << "WorkgroupMemoryExplicitLayout16BitAccessKHR" << ", " << "MultiView" << ", " << "VariablePointersStorageBuffer" << ", " << "RayQueryProvisionalKHR" << ", " << "RayQueryKHR" << ", " << "RayTracingKHR" << ", " << "Float16ImageAMD" << ", " << "ImageGatherBiasLodAMD" << ", " << "FragmentMaskAMD" << ", " << "StencilExportEXT" << ", " << "ImageReadWriteLodAMD" << ", " << "Int64ImageEXT" << ", " << "ShaderClockKHR" << ", " << "FragmentFullyCoveredEXT" << ", " << "MeshShadingNV" << ", " << "FragmentDensityEXT" << ", " << "ShaderNonUniform" << ", " << "RuntimeDescriptorArray" << ", " << "StorageTexelBufferArrayDynamicIndexing" << ", " << "RayTracingNV" << ", " << "RayTracingMotionBlurNV" << ", " << "PhysicalStorageBufferAddresses" << ", " << "RayTracingProvisionalKHR" << ", " << "CooperativeMatrixNV" << ", " << "FragmentShaderSampleInterlockEXT" << ", " << "FragmentShaderShadingRateInterlockEXT" << ", " << "ShaderSMBuiltinsNV" << ", " << "FragmentShaderPixelInterlockEXT" << ", " << "DemoteToHelperInvocation" << ", " << "IntegerFunctions2INTEL" << ", " << "TessellationPointSize" << ", " << "GeometryPointSize" << ", " << "ImageCubeArray" << ", " << "ImageRect" << ", " << "GeometryStreams" << ", " << "MultiViewport" << ", " << "WorkgroupMemoryExplicitLayout8BitAccessKHR" << ", " << "VariablePointers" << ", " << "RayTraversalPrimitiveCullingKHR" << ", " << "SampleMaskOverrideCoverageNV" << ", " << "GeometryShaderPassthroughNV" << ", " << "PerViewAttributesNV" << ", " << "InputAttachmentArrayDynamicIndexing" << ", " << "UniformBufferArrayNonUniformIndexing" << ", " << "SampledImageArrayNonUniformIndexing" << ", " << "StorageBufferArrayNonUniformIndexing" << ", " << "StorageImageArrayNonUniformIndexing" << ", " << "InputAttachmentArrayNonUniformIndexing" << ", " << "UniformTexelBufferArrayNonUniformIndexing" << ", " << "StorageTexelBufferArrayNonUniformIndexing" << ", " << "ShaderViewportIndexLayerEXT" << ", " << "ShaderViewportMaskNV" << ", " << "ShaderStereoViewNV" << ", " << "JointMatrixINTEL")};
634 }();
635 if (::mlir::failed(_result_value)) {
636 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CapabilityAttr parameter 'value' which is to be a `::mlir::spirv::Capability`");
637 return {};
638 }
639 // Parse literal '>'
640 if (odsParser.parseGreater()) return {};
641 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 641, __extension__ __PRETTY_FUNCTION__));
642 return CapabilityAttr::get(odsParser.getContext(),
643 ::mlir::spirv::Capability((*_result_value)));
644}
645
646void CapabilityAttr::print(::mlir::AsmPrinter &odsPrinter) const {
647 ::mlir::Builder odsBuilder(getContext());
648 odsPrinter << "<";
649 odsPrinter << stringifyCapability(getValue());
650 odsPrinter << ">";
651}
652
653::mlir::spirv::Capability CapabilityAttr::getValue() const {
654 return getImpl()->value;
655}
656
657} // namespace spirv
658} // namespace mlir
659MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::CapabilityAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::CapabilityAttr>::id = {}; } }
660namespace mlir {
661namespace spirv {
662namespace detail {
663struct CooperativeMatrixPropertiesNVAttrStorage : public ::mlir::AttributeStorage {
664 using KeyTy = std::tuple<int, int, int, mlir::Type, mlir::Type, mlir::Type, mlir::Type, mlir::spirv::ScopeAttr>;
665 CooperativeMatrixPropertiesNVAttrStorage(int m_size, int n_size, int k_size, mlir::Type a_type, mlir::Type b_type, mlir::Type c_type, mlir::Type result_type, mlir::spirv::ScopeAttr scope) : m_size(m_size), n_size(n_size), k_size(k_size), a_type(a_type), b_type(b_type), c_type(c_type), result_type(result_type), scope(scope) {}
666
667 KeyTy getAsKey() const {
668 return KeyTy(m_size, n_size, k_size, a_type, b_type, c_type, result_type, scope);
669 }
670
671 bool operator==(const KeyTy &tblgenKey) const {
672 return (m_size == std::get<0>(tblgenKey)) && (n_size == std::get<1>(tblgenKey)) && (k_size == std::get<2>(tblgenKey)) && (a_type == std::get<3>(tblgenKey)) && (b_type == std::get<4>(tblgenKey)) && (c_type == std::get<5>(tblgenKey)) && (result_type == std::get<6>(tblgenKey)) && (scope == std::get<7>(tblgenKey));
673 }
674
675 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
676 return ::llvm::hash_combine(std::get<0>(tblgenKey), std::get<1>(tblgenKey), std::get<2>(tblgenKey), std::get<3>(tblgenKey), std::get<4>(tblgenKey), std::get<5>(tblgenKey), std::get<6>(tblgenKey), std::get<7>(tblgenKey));
677 }
678
679 static CooperativeMatrixPropertiesNVAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
680 auto m_size = std::get<0>(tblgenKey);
681 auto n_size = std::get<1>(tblgenKey);
682 auto k_size = std::get<2>(tblgenKey);
683 auto a_type = std::get<3>(tblgenKey);
684 auto b_type = std::get<4>(tblgenKey);
685 auto c_type = std::get<5>(tblgenKey);
686 auto result_type = std::get<6>(tblgenKey);
687 auto scope = std::get<7>(tblgenKey);
688 return new (allocator.allocate<CooperativeMatrixPropertiesNVAttrStorage>()) CooperativeMatrixPropertiesNVAttrStorage(m_size, n_size, k_size, a_type, b_type, c_type, result_type, scope);
689 }
690
691 int m_size;
692 int n_size;
693 int k_size;
694 mlir::Type a_type;
695 mlir::Type b_type;
696 mlir::Type c_type;
697 mlir::Type result_type;
698 mlir::spirv::ScopeAttr scope;
699};
700} // namespace detail
701CooperativeMatrixPropertiesNVAttr CooperativeMatrixPropertiesNVAttr::get(::mlir::MLIRContext *context, int m_size, int n_size, int k_size, mlir::Type a_type, mlir::Type b_type, mlir::Type c_type, mlir::Type result_type, mlir::spirv::ScopeAttr scope) {
702 return Base::get(context, m_size, n_size, k_size, a_type, b_type, c_type, result_type, scope);
703}
704
705::mlir::Attribute CooperativeMatrixPropertiesNVAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
706 ::mlir::Builder odsBuilder(odsParser.getContext());
707 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
708 (void) odsLoc;
709 ::mlir::FailureOr<int> _result_m_size;
710 ::mlir::FailureOr<int> _result_n_size;
711 ::mlir::FailureOr<int> _result_k_size;
712 ::mlir::FailureOr<mlir::Type> _result_a_type;
713 ::mlir::FailureOr<mlir::Type> _result_b_type;
714 ::mlir::FailureOr<mlir::Type> _result_c_type;
715 ::mlir::FailureOr<mlir::Type> _result_result_type;
716 ::mlir::FailureOr<mlir::spirv::ScopeAttr> _result_scope;
717 // Parse literal '<'
718 if (odsParser.parseLess()) return {};
719 // Parse parameter struct
720 bool _seen_m_size = false;
721 bool _seen_n_size = false;
722 bool _seen_k_size = false;
723 bool _seen_a_type = false;
724 bool _seen_b_type = false;
725 bool _seen_c_type = false;
726 bool _seen_result_type = false;
727 bool _seen_scope = false;
728 {
729 const auto _loop_body = [&](::llvm::StringRef _paramKey) -> bool {
730 // Parse literal '='
731 if (odsParser.parseEqual()) return {};
732 if (!_seen_m_size && _paramKey == "m_size") {
733 _seen_m_size = true;
734
735 // Parse variable 'm_size'
736 _result_m_size = ::mlir::FieldParser<int>::parse(odsParser);
737 if (::mlir::failed(_result_m_size)) {
738 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'm_size' which is to be a `int`");
739 return {};
740 }
741 } else if (!_seen_n_size && _paramKey == "n_size") {
742 _seen_n_size = true;
743
744 // Parse variable 'n_size'
745 _result_n_size = ::mlir::FieldParser<int>::parse(odsParser);
746 if (::mlir::failed(_result_n_size)) {
747 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'n_size' which is to be a `int`");
748 return {};
749 }
750 } else if (!_seen_k_size && _paramKey == "k_size") {
751 _seen_k_size = true;
752
753 // Parse variable 'k_size'
754 _result_k_size = ::mlir::FieldParser<int>::parse(odsParser);
755 if (::mlir::failed(_result_k_size)) {
756 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'k_size' which is to be a `int`");
757 return {};
758 }
759 } else if (!_seen_a_type && _paramKey == "a_type") {
760 _seen_a_type = true;
761
762 // Parse variable 'a_type'
763 _result_a_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
764 if (::mlir::failed(_result_a_type)) {
765 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'a_type' which is to be a `mlir::Type`");
766 return {};
767 }
768 } else if (!_seen_b_type && _paramKey == "b_type") {
769 _seen_b_type = true;
770
771 // Parse variable 'b_type'
772 _result_b_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
773 if (::mlir::failed(_result_b_type)) {
774 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'b_type' which is to be a `mlir::Type`");
775 return {};
776 }
777 } else if (!_seen_c_type && _paramKey == "c_type") {
778 _seen_c_type = true;
779
780 // Parse variable 'c_type'
781 _result_c_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
782 if (::mlir::failed(_result_c_type)) {
783 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'c_type' which is to be a `mlir::Type`");
784 return {};
785 }
786 } else if (!_seen_result_type && _paramKey == "result_type") {
787 _seen_result_type = true;
788
789 // Parse variable 'result_type'
790 _result_result_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
791 if (::mlir::failed(_result_result_type)) {
792 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'result_type' which is to be a `mlir::Type`");
793 return {};
794 }
795 } else if (!_seen_scope && _paramKey == "scope") {
796 _seen_scope = true;
797
798 // Parse variable 'scope'
799 _result_scope = ::mlir::FieldParser<mlir::spirv::ScopeAttr>::parse(odsParser);
800 if (::mlir::failed(_result_scope)) {
801 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'scope' which is to be a `mlir::spirv::ScopeAttr`");
802 return {};
803 }
804 } else {
805 odsParser.emitError(odsParser.getCurrentLocation(), "duplicate or unknown struct parameter name: ") << _paramKey;
806 return {};
807 }
808 return true;
809 };
810 for (unsigned odsStructIndex = 0; odsStructIndex < 8; ++odsStructIndex) {
811 ::llvm::StringRef _paramKey;
812 if (odsParser.parseKeyword(&_paramKey)) {
813 odsParser.emitError(odsParser.getCurrentLocation(),
814 "expected a parameter name in struct");
815 return {};
816 }
817 if (!_loop_body(_paramKey)) return {};
818 if ((odsStructIndex != 8 - 1) && odsParser.parseComma())
819 return {};
820 }
821 }
822 // Parse literal '>'
823 if (odsParser.parseGreater()) return {};
824 assert(::mlir::succeeded(_result_m_size))(static_cast <bool> (::mlir::succeeded(_result_m_size))
? void (0) : __assert_fail ("::mlir::succeeded(_result_m_size)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 824, __extension__ __PRETTY_FUNCTION__));
825 assert(::mlir::succeeded(_result_n_size))(static_cast <bool> (::mlir::succeeded(_result_n_size))
? void (0) : __assert_fail ("::mlir::succeeded(_result_n_size)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 825, __extension__ __PRETTY_FUNCTION__));
826 assert(::mlir::succeeded(_result_k_size))(static_cast <bool> (::mlir::succeeded(_result_k_size))
? void (0) : __assert_fail ("::mlir::succeeded(_result_k_size)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 826, __extension__ __PRETTY_FUNCTION__));
827 assert(::mlir::succeeded(_result_a_type))(static_cast <bool> (::mlir::succeeded(_result_a_type))
? void (0) : __assert_fail ("::mlir::succeeded(_result_a_type)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 827, __extension__ __PRETTY_FUNCTION__));
828 assert(::mlir::succeeded(_result_b_type))(static_cast <bool> (::mlir::succeeded(_result_b_type))
? void (0) : __assert_fail ("::mlir::succeeded(_result_b_type)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 828, __extension__ __PRETTY_FUNCTION__));
829 assert(::mlir::succeeded(_result_c_type))(static_cast <bool> (::mlir::succeeded(_result_c_type))
? void (0) : __assert_fail ("::mlir::succeeded(_result_c_type)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 829, __extension__ __PRETTY_FUNCTION__));
830 assert(::mlir::succeeded(_result_result_type))(static_cast <bool> (::mlir::succeeded(_result_result_type
)) ? void (0) : __assert_fail ("::mlir::succeeded(_result_result_type)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 830, __extension__ __PRETTY_FUNCTION__));
831 assert(::mlir::succeeded(_result_scope))(static_cast <bool> (::mlir::succeeded(_result_scope)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_scope)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 831, __extension__ __PRETTY_FUNCTION__));
832 return CooperativeMatrixPropertiesNVAttr::get(odsParser.getContext(),
833 int((*_result_m_size)),
834 int((*_result_n_size)),
835 int((*_result_k_size)),
836 mlir::Type((*_result_a_type)),
837 mlir::Type((*_result_b_type)),
838 mlir::Type((*_result_c_type)),
839 mlir::Type((*_result_result_type)),
840 mlir::spirv::ScopeAttr((*_result_scope)));
841}
842
843void CooperativeMatrixPropertiesNVAttr::print(::mlir::AsmPrinter &odsPrinter) const {
844 ::mlir::Builder odsBuilder(getContext());
845 odsPrinter << "<";
846 {
847 bool _firstPrinted = true;
848 if (!_firstPrinted) odsPrinter << ", ";
849 _firstPrinted = false;
850 odsPrinter << "m_size = ";
851 odsPrinter.printStrippedAttrOrType(getMSize());
852 if (!_firstPrinted) odsPrinter << ", ";
853 _firstPrinted = false;
854 odsPrinter << "n_size = ";
855 odsPrinter.printStrippedAttrOrType(getNSize());
856 if (!_firstPrinted) odsPrinter << ", ";
857 _firstPrinted = false;
858 odsPrinter << "k_size = ";
859 odsPrinter.printStrippedAttrOrType(getKSize());
860 if (!_firstPrinted) odsPrinter << ", ";
861 _firstPrinted = false;
862 odsPrinter << "a_type = ";
863 odsPrinter.printStrippedAttrOrType(getAType());
864 if (!_firstPrinted) odsPrinter << ", ";
865 _firstPrinted = false;
866 odsPrinter << "b_type = ";
867 odsPrinter.printStrippedAttrOrType(getBType());
868 if (!_firstPrinted) odsPrinter << ", ";
869 _firstPrinted = false;
870 odsPrinter << "c_type = ";
871 odsPrinter.printStrippedAttrOrType(getCType());
872 if (!_firstPrinted) odsPrinter << ", ";
873 _firstPrinted = false;
874 odsPrinter << "result_type = ";
875 odsPrinter.printStrippedAttrOrType(getResultType());
876 if (!_firstPrinted) odsPrinter << ", ";
877 _firstPrinted = false;
878 odsPrinter << "scope = ";
879 odsPrinter.printStrippedAttrOrType(getScope());
880 }
881 odsPrinter << ">";
882}
883
884int CooperativeMatrixPropertiesNVAttr::getMSize() const {
885 return getImpl()->m_size;
886}
887
888int CooperativeMatrixPropertiesNVAttr::getNSize() const {
889 return getImpl()->n_size;
890}
891
892int CooperativeMatrixPropertiesNVAttr::getKSize() const {
893 return getImpl()->k_size;
894}
895
896mlir::Type CooperativeMatrixPropertiesNVAttr::getAType() const {
897 return getImpl()->a_type;
898}
899
900mlir::Type CooperativeMatrixPropertiesNVAttr::getBType() const {
901 return getImpl()->b_type;
902}
903
904mlir::Type CooperativeMatrixPropertiesNVAttr::getCType() const {
905 return getImpl()->c_type;
906}
907
908mlir::Type CooperativeMatrixPropertiesNVAttr::getResultType() const {
909 return getImpl()->result_type;
910}
911
912mlir::spirv::ScopeAttr CooperativeMatrixPropertiesNVAttr::getScope() const {
913 return getImpl()->scope;
914}
915
916} // namespace spirv
917} // namespace mlir
918MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::CooperativeMatrixPropertiesNVAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::CooperativeMatrixPropertiesNVAttr>::id
= {}; } }
919namespace mlir {
920namespace spirv {
921namespace detail {
922struct DecorationAttrStorage : public ::mlir::AttributeStorage {
923 using KeyTy = std::tuple<::mlir::spirv::Decoration>;
924 DecorationAttrStorage(::mlir::spirv::Decoration value) : value(value) {}
925
926 KeyTy getAsKey() const {
927 return KeyTy(value);
928 }
929
930 bool operator==(const KeyTy &tblgenKey) const {
931 return (value == std::get<0>(tblgenKey));
932 }
933
934 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
935 return ::llvm::hash_combine(std::get<0>(tblgenKey));
936 }
937
938 static DecorationAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
939 auto value = std::get<0>(tblgenKey);
940 return new (allocator.allocate<DecorationAttrStorage>()) DecorationAttrStorage(value);
941 }
942
943 ::mlir::spirv::Decoration value;
944};
945} // namespace detail
946DecorationAttr DecorationAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Decoration value) {
947 return Base::get(context, value);
948}
949
950::mlir::Attribute DecorationAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
951 ::mlir::Builder odsBuilder(odsParser.getContext());
952 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
953 (void) odsLoc;
954 ::mlir::FailureOr<::mlir::spirv::Decoration> _result_value;
955 // Parse literal '<'
956 if (odsParser.parseLess()) return {};
957
958 // Parse variable 'value'
959 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Decoration> {
960 auto loc = odsParser.getCurrentLocation();
961 ::llvm::StringRef enumKeyword;
962 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
963 return ::mlir::failure();
964 auto maybeEnum = ::mlir::spirv::symbolizeDecoration(enumKeyword);
965 if (maybeEnum)
966 return *maybeEnum;
967 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::Decoration" << " to be one of: " << "RelaxedPrecision" << ", " << "SpecId" << ", " << "Block" << ", " << "BufferBlock" << ", " << "RowMajor" << ", " << "ColMajor" << ", " << "ArrayStride" << ", " << "MatrixStride" << ", " << "GLSLShared" << ", " << "GLSLPacked" << ", " << "CPacked" << ", " << "BuiltIn" << ", " << "NoPerspective" << ", " << "Flat" << ", " << "Patch" << ", " << "Centroid" << ", " << "Sample" << ", " << "Invariant" << ", " << "Restrict" << ", " << "Aliased" << ", " << "Volatile" << ", " << "Constant" << ", " << "Coherent" << ", " << "NonWritable" << ", " << "NonReadable" << ", " << "Uniform" << ", " << "UniformId" << ", " << "SaturatedConversion" << ", " << "Stream" << ", " << "Location" << ", " << "Component" << ", " << "Index" << ", " << "Binding" << ", " << "DescriptorSet" << ", " << "Offset" << ", " << "XfbBuffer" << ", " << "XfbStride" << ", " << "FuncParamAttr" << ", " << "FPRoundingMode" << ", " << "FPFastMathMode" << ", " << "LinkageAttributes" << ", " << "NoContraction" << ", " << "InputAttachmentIndex" << ", " << "Alignment" << ", " << "MaxByteOffset" << ", " << "AlignmentId" << ", " << "MaxByteOffsetId" << ", " << "NoSignedWrap" << ", " << "NoUnsignedWrap" << ", " << "ExplicitInterpAMD" << ", " << "OverrideCoverageNV" << ", " << "PassthroughNV" << ", " << "ViewportRelativeNV" << ", " << "SecondaryViewportRelativeNV" << ", " << "PerPrimitiveNV" << ", " << "PerViewNV" << ", " << "PerTaskNV" << ", " << "PerVertexKHR" << ", " << "NonUniform" << ", " << "RestrictPointer" << ", " << "AliasedPointer" << ", " << "BindlessSamplerNV" << ", " << "BindlessImageNV" << ", " << "BoundSamplerNV" << ", " << "BoundImageNV" << ", " << "SIMTCallINTEL" << ", " << "ReferencedIndirectlyINTEL" << ", " << "ClobberINTEL" << ", " << "SideEffectsINTEL" << ", " << "VectorComputeVariableINTEL" << ", " << "FuncParamIOKindINTEL" << ", " << "VectorComputeFunctionINTEL" << ", " << "StackCallINTEL" << ", " << "GlobalVariableOffsetINTEL" << ", " << "CounterBuffer" << ", " << "UserSemantic" << ", " << "UserTypeGOOGLE" << ", " << "FunctionRoundingModeINTEL" << ", " << "FunctionDenormModeINTEL" << ", " << "RegisterINTEL" << ", " << "MemoryINTEL" << ", " << "NumbanksINTEL" << ", " << "BankwidthINTEL" << ", " << "MaxPrivateCopiesINTEL" << ", " << "SinglepumpINTEL" << ", " << "DoublepumpINTEL" << ", " << "MaxReplicatesINTEL" << ", " << "SimpleDualPortINTEL" << ", " << "MergeINTEL" << ", " << "BankBitsINTEL" << ", " << "ForcePow2DepthINTEL" << ", " << "BurstCoalesceINTEL" << ", " << "CacheSizeINTEL" << ", " << "DontStaticallyCoalesceINTEL" << ", " << "PrefetchINTEL" << ", " << "StallEnableINTEL" << ", " << "FuseLoopsInFunctionINTEL" << ", " << "AliasScopeINTEL" << ", " << "NoAliasINTEL" << ", " << "BufferLocationINTEL" << ", " << "IOPipeStorageINTEL" << ", " << "FunctionFloatingPointModeINTEL" << ", " << "SingleElementVectorINTEL" << ", " << "VectorComputeCallableFunctionINTEL" << ", " << "MediaBlockIOINTEL")};
968 }();
969 if (::mlir::failed(_result_value)) {
970 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_DecorationAttr parameter 'value' which is to be a `::mlir::spirv::Decoration`");
971 return {};
972 }
973 // Parse literal '>'
974 if (odsParser.parseGreater()) return {};
975 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 975, __extension__ __PRETTY_FUNCTION__));
976 return DecorationAttr::get(odsParser.getContext(),
977 ::mlir::spirv::Decoration((*_result_value)));
978}
979
980void DecorationAttr::print(::mlir::AsmPrinter &odsPrinter) const {
981 ::mlir::Builder odsBuilder(getContext());
982 odsPrinter << "<";
983 odsPrinter << stringifyDecoration(getValue());
984 odsPrinter << ">";
985}
986
987::mlir::spirv::Decoration DecorationAttr::getValue() const {
988 return getImpl()->value;
989}
990
991} // namespace spirv
992} // namespace mlir
993MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::DecorationAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::DecorationAttr>::id = {}; } }
994namespace mlir {
995namespace spirv {
996namespace detail {
997struct ImageDepthInfoAttrStorage : public ::mlir::AttributeStorage {
998 using KeyTy = std::tuple<::mlir::spirv::ImageDepthInfo>;
999 ImageDepthInfoAttrStorage(::mlir::spirv::ImageDepthInfo value) : value(value) {}
1000
1001 KeyTy getAsKey() const {
1002 return KeyTy(value);
1003 }
1004
1005 bool operator==(const KeyTy &tblgenKey) const {
1006 return (value == std::get<0>(tblgenKey));
1007 }
1008
1009 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1010 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1011 }
1012
1013 static ImageDepthInfoAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1014 auto value = std::get<0>(tblgenKey);
1015 return new (allocator.allocate<ImageDepthInfoAttrStorage>()) ImageDepthInfoAttrStorage(value);
1016 }
1017
1018 ::mlir::spirv::ImageDepthInfo value;
1019};
1020} // namespace detail
1021ImageDepthInfoAttr ImageDepthInfoAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ImageDepthInfo value) {
1022 return Base::get(context, value);
1023}
1024
1025::mlir::Attribute ImageDepthInfoAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1026 ::mlir::Builder odsBuilder(odsParser.getContext());
1027 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1028 (void) odsLoc;
1029 ::mlir::FailureOr<::mlir::spirv::ImageDepthInfo> _result_value;
1030 // Parse literal '<'
1031 if (odsParser.parseLess()) return {};
1032
1033 // Parse variable 'value'
1034 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ImageDepthInfo> {
1035 auto loc = odsParser.getCurrentLocation();
1036 ::llvm::StringRef enumKeyword;
1037 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1038 return ::mlir::failure();
1039 auto maybeEnum = ::mlir::spirv::symbolizeImageDepthInfo(enumKeyword);
1040 if (maybeEnum)
1041 return *maybeEnum;
1042 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ImageDepthInfo" << " to be one of: " << "NoDepth" << ", " << "IsDepth" << ", " << "DepthUnknown")};
1043 }();
1044 if (::mlir::failed(_result_value)) {
1045 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_DepthAttr parameter 'value' which is to be a `::mlir::spirv::ImageDepthInfo`");
1046 return {};
1047 }
1048 // Parse literal '>'
1049 if (odsParser.parseGreater()) return {};
1050 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 1050, __extension__ __PRETTY_FUNCTION__));
1051 return ImageDepthInfoAttr::get(odsParser.getContext(),
1052 ::mlir::spirv::ImageDepthInfo((*_result_value)));
1053}
1054
1055void ImageDepthInfoAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1056 ::mlir::Builder odsBuilder(getContext());
1057 odsPrinter << "<";
1058 odsPrinter << stringifyImageDepthInfo(getValue());
1059 odsPrinter << ">";
1060}
1061
1062::mlir::spirv::ImageDepthInfo ImageDepthInfoAttr::getValue() const {
1063 return getImpl()->value;
1064}
1065
1066} // namespace spirv
1067} // namespace mlir
1068MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ImageDepthInfoAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ImageDepthInfoAttr>::id = {}; } }
1069namespace mlir {
1070namespace spirv {
1071namespace detail {
1072struct DeviceTypeAttrStorage : public ::mlir::AttributeStorage {
1073 using KeyTy = std::tuple<::mlir::spirv::DeviceType>;
1074 DeviceTypeAttrStorage(::mlir::spirv::DeviceType value) : value(value) {}
1075
1076 KeyTy getAsKey() const {
1077 return KeyTy(value);
1078 }
1079
1080 bool operator==(const KeyTy &tblgenKey) const {
1081 return (value == std::get<0>(tblgenKey));
1082 }
1083
1084 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1085 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1086 }
1087
1088 static DeviceTypeAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1089 auto value = std::get<0>(tblgenKey);
1090 return new (allocator.allocate<DeviceTypeAttrStorage>()) DeviceTypeAttrStorage(value);
1091 }
1092
1093 ::mlir::spirv::DeviceType value;
1094};
1095} // namespace detail
1096DeviceTypeAttr DeviceTypeAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::DeviceType value) {
1097 return Base::get(context, value);
1098}
1099
1100::mlir::Attribute DeviceTypeAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1101 ::mlir::Builder odsBuilder(odsParser.getContext());
1102 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1103 (void) odsLoc;
1104 ::mlir::FailureOr<::mlir::spirv::DeviceType> _result_value;
1105 // Parse literal '<'
1106 if (odsParser.parseLess()) return {};
1107
1108 // Parse variable 'value'
1109 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::DeviceType> {
1110 auto loc = odsParser.getCurrentLocation();
1111 ::llvm::StringRef enumKeyword;
1112 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1113 return ::mlir::failure();
1114 auto maybeEnum = ::mlir::spirv::symbolizeDeviceType(enumKeyword);
1115 if (maybeEnum)
1116 return *maybeEnum;
1117 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::DeviceType" << " to be one of: " << "Other" << ", " << "IntegratedGPU" << ", " << "DiscreteGPU" << ", " << "CPU" << ", " << "Unknown")};
1118 }();
1119 if (::mlir::failed(_result_value)) {
1120 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_DeviceTypeAttr parameter 'value' which is to be a `::mlir::spirv::DeviceType`");
1121 return {};
1122 }
1123 // Parse literal '>'
1124 if (odsParser.parseGreater()) return {};
1125 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 1125, __extension__ __PRETTY_FUNCTION__));
1126 return DeviceTypeAttr::get(odsParser.getContext(),
1127 ::mlir::spirv::DeviceType((*_result_value)));
1128}
1129
1130void DeviceTypeAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1131 ::mlir::Builder odsBuilder(getContext());
1132 odsPrinter << "<";
1133 odsPrinter << stringifyDeviceType(getValue());
1134 odsPrinter << ">";
1135}
1136
1137::mlir::spirv::DeviceType DeviceTypeAttr::getValue() const {
1138 return getImpl()->value;
1139}
1140
1141} // namespace spirv
1142} // namespace mlir
1143MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::DeviceTypeAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::DeviceTypeAttr>::id = {}; } }
1144namespace mlir {
1145namespace spirv {
1146namespace detail {
1147struct DimAttrStorage : public ::mlir::AttributeStorage {
1148 using KeyTy = std::tuple<::mlir::spirv::Dim>;
1149 DimAttrStorage(::mlir::spirv::Dim value) : value(value) {}
1150
1151 KeyTy getAsKey() const {
1152 return KeyTy(value);
1153 }
1154
1155 bool operator==(const KeyTy &tblgenKey) const {
1156 return (value == std::get<0>(tblgenKey));
1157 }
1158
1159 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1160 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1161 }
1162
1163 static DimAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1164 auto value = std::get<0>(tblgenKey);
1165 return new (allocator.allocate<DimAttrStorage>()) DimAttrStorage(value);
1166 }
1167
1168 ::mlir::spirv::Dim value;
1169};
1170} // namespace detail
1171DimAttr DimAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Dim value) {
1172 return Base::get(context, value);
1173}
1174
1175::mlir::Attribute DimAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1176 ::mlir::Builder odsBuilder(odsParser.getContext());
1177 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1178 (void) odsLoc;
1179 ::mlir::FailureOr<::mlir::spirv::Dim> _result_value;
1180 // Parse literal '<'
1181 if (odsParser.parseLess()) return {};
1182
1183 // Parse variable 'value'
1184 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Dim> {
1185 auto loc = odsParser.getCurrentLocation();
1186 ::llvm::StringRef enumKeyword;
1187 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1188 return ::mlir::failure();
1189 auto maybeEnum = ::mlir::spirv::symbolizeDim(enumKeyword);
1190 if (maybeEnum)
1191 return *maybeEnum;
1192 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::Dim" << " to be one of: " << "Dim1D" << ", " << "Dim2D" << ", " << "Dim3D" << ", " << "Cube" << ", " << "Rect" << ", " << "Buffer" << ", " << "SubpassData")};
1193 }();
1194 if (::mlir::failed(_result_value)) {
1195 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_DimAttr parameter 'value' which is to be a `::mlir::spirv::Dim`");
1196 return {};
1197 }
1198 // Parse literal '>'
1199 if (odsParser.parseGreater()) return {};
1200 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 1200, __extension__ __PRETTY_FUNCTION__));
1201 return DimAttr::get(odsParser.getContext(),
1202 ::mlir::spirv::Dim((*_result_value)));
1203}
1204
1205void DimAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1206 ::mlir::Builder odsBuilder(getContext());
1207 odsPrinter << "<";
1208 odsPrinter << stringifyDim(getValue());
1209 odsPrinter << ">";
1210}
1211
1212::mlir::spirv::Dim DimAttr::getValue() const {
1213 return getImpl()->value;
1214}
1215
1216} // namespace spirv
1217} // namespace mlir
1218MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::DimAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::DimAttr>::id = {}; } }
1219namespace mlir {
1220namespace spirv {
1221namespace detail {
1222struct EntryPointABIAttrStorage : public ::mlir::AttributeStorage {
1223 using KeyTy = std::tuple<DenseIntElementsAttr>;
1224 EntryPointABIAttrStorage(DenseIntElementsAttr local_size) : local_size(local_size) {}
1225
1226 KeyTy getAsKey() const {
1227 return KeyTy(local_size);
1228 }
1229
1230 bool operator==(const KeyTy &tblgenKey) const {
1231 return (local_size == std::get<0>(tblgenKey));
1232 }
1233
1234 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1235 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1236 }
1237
1238 static EntryPointABIAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1239 auto local_size = std::get<0>(tblgenKey);
1240 return new (allocator.allocate<EntryPointABIAttrStorage>()) EntryPointABIAttrStorage(local_size);
1241 }
1242
1243 DenseIntElementsAttr local_size;
1244};
1245} // namespace detail
1246EntryPointABIAttr EntryPointABIAttr::get(::mlir::MLIRContext *context, DenseIntElementsAttr local_size) {
1247 return Base::get(context, local_size);
1248}
1249
1250::mlir::Attribute EntryPointABIAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1251 ::mlir::Builder odsBuilder(odsParser.getContext());
1252 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1253 (void) odsLoc;
1254 ::mlir::FailureOr<DenseIntElementsAttr> _result_local_size;
1255 // Parse literal '<'
1256 if (odsParser.parseLess()) return {};
1257 // Parse parameter struct
1258 bool _seen_local_size = false;
1259 {
1260 const auto _loop_body = [&](::llvm::StringRef _paramKey) -> bool {
1261 // Parse literal '='
1262 if (odsParser.parseEqual()) return {};
1263 if (!_seen_local_size && _paramKey == "local_size") {
1264 _seen_local_size = true;
1265
1266 // Parse variable 'local_size'
1267 _result_local_size = ::mlir::FieldParser<DenseIntElementsAttr>::parse(odsParser);
1268 if (::mlir::failed(_result_local_size)) {
1269 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_EntryPointABIAttr parameter 'local_size' which is to be a `DenseIntElementsAttr`");
1270 return {};
1271 }
1272 } else {
1273 odsParser.emitError(odsParser.getCurrentLocation(), "duplicate or unknown struct parameter name: ") << _paramKey;
1274 return {};
1275 }
1276 return true;
1277 };
1278 ::llvm::StringRef _paramKey;
1279 if (!odsParser.parseOptionalKeyword(&_paramKey)) {
1280 if (!_loop_body(_paramKey)) return {};
1281 while (!odsParser.parseOptionalComma()) {
1282 ::llvm::StringRef _paramKey;
1283 if (odsParser.parseKeyword(&_paramKey)) {
1284 odsParser.emitError(odsParser.getCurrentLocation(),
1285 "expected a parameter name in struct");
1286 return {};
1287 }
1288 if (!_loop_body(_paramKey)) return {};
1289 }
1290 }
1291 }
1292 // Parse literal '>'
1293 if (odsParser.parseGreater()) return {};
1294 return EntryPointABIAttr::get(odsParser.getContext(),
1295 DenseIntElementsAttr((_result_local_size.value_or(DenseIntElementsAttr()))));
1296}
1297
1298void EntryPointABIAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1299 ::mlir::Builder odsBuilder(getContext());
1300 odsPrinter << "<";
1301 {
1302 bool _firstPrinted = true;
1303 if (!(getLocalSize() == DenseIntElementsAttr())) {
1304 if (!_firstPrinted) odsPrinter << ", ";
1305 _firstPrinted = false;
1306 odsPrinter << "local_size = ";
1307 if (!(getLocalSize() == DenseIntElementsAttr())) {
1308 odsPrinter.printStrippedAttrOrType(getLocalSize());
1309 }
1310 }
1311 }
1312 odsPrinter << ">";
1313}
1314
1315DenseIntElementsAttr EntryPointABIAttr::getLocalSize() const {
1316 return getImpl()->local_size;
1317}
1318
1319} // namespace spirv
1320} // namespace mlir
1321MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::EntryPointABIAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::EntryPointABIAttr>::id = {}; } }
1322namespace mlir {
1323namespace spirv {
1324namespace detail {
1325struct ExecutionModeAttrStorage : public ::mlir::AttributeStorage {
1326 using KeyTy = std::tuple<::mlir::spirv::ExecutionMode>;
1327 ExecutionModeAttrStorage(::mlir::spirv::ExecutionMode value) : value(value) {}
1328
1329 KeyTy getAsKey() const {
1330 return KeyTy(value);
1331 }
1332
1333 bool operator==(const KeyTy &tblgenKey) const {
1334 return (value == std::get<0>(tblgenKey));
1335 }
1336
1337 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1338 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1339 }
1340
1341 static ExecutionModeAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1342 auto value = std::get<0>(tblgenKey);
1343 return new (allocator.allocate<ExecutionModeAttrStorage>()) ExecutionModeAttrStorage(value);
1344 }
1345
1346 ::mlir::spirv::ExecutionMode value;
1347};
1348} // namespace detail
1349ExecutionModeAttr ExecutionModeAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ExecutionMode value) {
1350 return Base::get(context, value);
1351}
1352
1353::mlir::Attribute ExecutionModeAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1354 ::mlir::Builder odsBuilder(odsParser.getContext());
1355 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1356 (void) odsLoc;
1357 ::mlir::FailureOr<::mlir::spirv::ExecutionMode> _result_value;
1358 // Parse literal '<'
1359 if (odsParser.parseLess()) return {};
1360
1361 // Parse variable 'value'
1362 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ExecutionMode> {
1363 auto loc = odsParser.getCurrentLocation();
1364 ::llvm::StringRef enumKeyword;
1365 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1366 return ::mlir::failure();
1367 auto maybeEnum = ::mlir::spirv::symbolizeExecutionMode(enumKeyword);
1368 if (maybeEnum)
1369 return *maybeEnum;
1370 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ExecutionMode" << " to be one of: " << "Invocations" << ", " << "SpacingEqual" << ", " << "SpacingFractionalEven" << ", " << "SpacingFractionalOdd" << ", " << "VertexOrderCw" << ", " << "VertexOrderCcw" << ", " << "PixelCenterInteger" << ", " << "OriginUpperLeft" << ", " << "OriginLowerLeft" << ", " << "EarlyFragmentTests" << ", " << "PointMode" << ", " << "Xfb" << ", " << "DepthReplacing" << ", " << "DepthGreater" << ", " << "DepthLess" << ", " << "DepthUnchanged" << ", " << "LocalSize" << ", " << "LocalSizeHint" << ", " << "InputPoints" << ", " << "InputLines" << ", " << "InputLinesAdjacency" << ", " << "Triangles" << ", " << "InputTrianglesAdjacency" << ", " << "Quads" << ", " << "Isolines" << ", " << "OutputVertices" << ", " << "OutputPoints" << ", " << "OutputLineStrip" << ", " << "OutputTriangleStrip" << ", " << "VecTypeHint" << ", " << "ContractionOff" << ", " << "Initializer" << ", " << "Finalizer" << ", " << "SubgroupSize" << ", " << "SubgroupsPerWorkgroup" << ", " << "SubgroupsPerWorkgroupId" << ", " << "LocalSizeId" << ", " << "LocalSizeHintId" << ", " << "SubgroupUniformControlFlowKHR" << ", " << "PostDepthCoverage" << ", " << "DenormPreserve" << ", " << "DenormFlushToZero" << ", " << "SignedZeroInfNanPreserve" << ", " << "RoundingModeRTE" << ", " << "RoundingModeRTZ" << ", " << "EarlyAndLateFragmentTestsAMD" << ", " << "StencilRefReplacingEXT" << ", " << "StencilRefUnchangedFrontAMD" << ", " << "StencilRefGreaterFrontAMD" << ", " << "StencilRefLessFrontAMD" << ", " << "StencilRefUnchangedBackAMD" << ", " << "StencilRefGreaterBackAMD" << ", " << "StencilRefLessBackAMD" << ", " << "OutputLinesNV" << ", " << "OutputPrimitivesNV" << ", " << "DerivativeGroupQuadsNV" << ", " << "DerivativeGroupLinearNV" << ", " << "OutputTrianglesNV" << ", " << "PixelInterlockOrderedEXT" << ", " << "PixelInterlockUnorderedEXT" << ", " << "SampleInterlockOrderedEXT" << ", " << "SampleInterlockUnorderedEXT" << ", " << "ShadingRateInterlockOrderedEXT" << ", " << "ShadingRateInterlockUnorderedEXT" << ", " << "SharedLocalMemorySizeINTEL" << ", " << "RoundingModeRTPINTEL" << ", " << "RoundingModeRTNINTEL" << ", " << "FloatingPointModeALTINTEL" << ", " << "FloatingPointModeIEEEINTEL" << ", " << "MaxWorkgroupSizeINTEL" << ", " << "MaxWorkDimINTEL" << ", " << "NoGlobalOffsetINTEL" << ", " << "NumSIMDWorkitemsINTEL" << ", " << "SchedulerTargetFmaxMhzINTEL" << ", " << "NamedBarrierCountINTEL")};
1371 }();
1372 if (::mlir::failed(_result_value)) {
1373 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ExecutionModeAttr parameter 'value' which is to be a `::mlir::spirv::ExecutionMode`");
1374 return {};
1375 }
1376 // Parse literal '>'
1377 if (odsParser.parseGreater()) return {};
1378 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 1378, __extension__ __PRETTY_FUNCTION__));
1379 return ExecutionModeAttr::get(odsParser.getContext(),
1380 ::mlir::spirv::ExecutionMode((*_result_value)));
1381}
1382
1383void ExecutionModeAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1384 ::mlir::Builder odsBuilder(getContext());
1385 odsPrinter << "<";
1386 odsPrinter << stringifyExecutionMode(getValue());
1387 odsPrinter << ">";
1388}
1389
1390::mlir::spirv::ExecutionMode ExecutionModeAttr::getValue() const {
1391 return getImpl()->value;
1392}
1393
1394} // namespace spirv
1395} // namespace mlir
1396MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ExecutionModeAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ExecutionModeAttr>::id = {}; } }
1397namespace mlir {
1398namespace spirv {
1399namespace detail {
1400struct ExecutionModelAttrStorage : public ::mlir::AttributeStorage {
1401 using KeyTy = std::tuple<::mlir::spirv::ExecutionModel>;
1402 ExecutionModelAttrStorage(::mlir::spirv::ExecutionModel value) : value(value) {}
1403
1404 KeyTy getAsKey() const {
1405 return KeyTy(value);
1406 }
1407
1408 bool operator==(const KeyTy &tblgenKey) const {
1409 return (value == std::get<0>(tblgenKey));
1410 }
1411
1412 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1413 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1414 }
1415
1416 static ExecutionModelAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1417 auto value = std::get<0>(tblgenKey);
1418 return new (allocator.allocate<ExecutionModelAttrStorage>()) ExecutionModelAttrStorage(value);
1419 }
1420
1421 ::mlir::spirv::ExecutionModel value;
1422};
1423} // namespace detail
1424ExecutionModelAttr ExecutionModelAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ExecutionModel value) {
1425 return Base::get(context, value);
1426}
1427
1428::mlir::Attribute ExecutionModelAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1429 ::mlir::Builder odsBuilder(odsParser.getContext());
1430 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1431 (void) odsLoc;
1432 ::mlir::FailureOr<::mlir::spirv::ExecutionModel> _result_value;
1433 // Parse literal '<'
1434 if (odsParser.parseLess()) return {};
1435
1436 // Parse variable 'value'
1437 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ExecutionModel> {
1438 auto loc = odsParser.getCurrentLocation();
1439 ::llvm::StringRef enumKeyword;
1440 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1441 return ::mlir::failure();
1442 auto maybeEnum = ::mlir::spirv::symbolizeExecutionModel(enumKeyword);
1443 if (maybeEnum)
1444 return *maybeEnum;
1445 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ExecutionModel" << " to be one of: " << "Vertex" << ", " << "TessellationControl" << ", " << "TessellationEvaluation" << ", " << "Geometry" << ", " << "Fragment" << ", " << "GLCompute" << ", " << "Kernel" << ", " << "TaskNV" << ", " << "MeshNV" << ", " << "RayGenerationKHR" << ", " << "IntersectionKHR" << ", " << "AnyHitKHR" << ", " << "ClosestHitKHR" << ", " << "MissKHR" << ", " << "CallableKHR")};
1446 }();
1447 if (::mlir::failed(_result_value)) {
1448 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ExecutionModelAttr parameter 'value' which is to be a `::mlir::spirv::ExecutionModel`");
1449 return {};
1450 }
1451 // Parse literal '>'
1452 if (odsParser.parseGreater()) return {};
1453 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 1453, __extension__ __PRETTY_FUNCTION__));
1454 return ExecutionModelAttr::get(odsParser.getContext(),
1455 ::mlir::spirv::ExecutionModel((*_result_value)));
1456}
1457
1458void ExecutionModelAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1459 ::mlir::Builder odsBuilder(getContext());
1460 odsPrinter << "<";
1461 odsPrinter << stringifyExecutionModel(getValue());
1462 odsPrinter << ">";
1463}
1464
1465::mlir::spirv::ExecutionModel ExecutionModelAttr::getValue() const {
1466 return getImpl()->value;
1467}
1468
1469} // namespace spirv
1470} // namespace mlir
1471MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ExecutionModelAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ExecutionModelAttr>::id = {}; } }
1472namespace mlir {
1473namespace spirv {
1474namespace detail {
1475struct ExtensionAttrStorage : public ::mlir::AttributeStorage {
1476 using KeyTy = std::tuple<::mlir::spirv::Extension>;
1477 ExtensionAttrStorage(::mlir::spirv::Extension value) : value(value) {}
1478
1479 KeyTy getAsKey() const {
1480 return KeyTy(value);
1481 }
1482
1483 bool operator==(const KeyTy &tblgenKey) const {
1484 return (value == std::get<0>(tblgenKey));
1485 }
1486
1487 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1488 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1489 }
1490
1491 static ExtensionAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1492 auto value = std::get<0>(tblgenKey);
1493 return new (allocator.allocate<ExtensionAttrStorage>()) ExtensionAttrStorage(value);
1494 }
1495
1496 ::mlir::spirv::Extension value;
1497};
1498} // namespace detail
1499ExtensionAttr ExtensionAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Extension value) {
1500 return Base::get(context, value);
1501}
1502
1503::mlir::Attribute ExtensionAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1504 ::mlir::Builder odsBuilder(odsParser.getContext());
1505 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1506 (void) odsLoc;
1507 ::mlir::FailureOr<::mlir::spirv::Extension> _result_value;
1508 // Parse literal '<'
1509 if (odsParser.parseLess()) return {};
1510
1511 // Parse variable 'value'
1512 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Extension> {
1513 auto loc = odsParser.getCurrentLocation();
1514 ::llvm::StringRef enumKeyword;
1515 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1516 return ::mlir::failure();
1517 auto maybeEnum = ::mlir::spirv::symbolizeExtension(enumKeyword);
1518 if (maybeEnum)
1519 return *maybeEnum;
1520 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::Extension" << " to be one of: " << "SPV_KHR_16bit_storage" << ", " << "SPV_KHR_8bit_storage" << ", " << "SPV_KHR_device_group" << ", " << "SPV_KHR_float_controls" << ", " << "SPV_KHR_physical_storage_buffer" << ", " << "SPV_KHR_multiview" << ", " << "SPV_KHR_no_integer_wrap_decoration" << ", " << "SPV_KHR_post_depth_coverage" << ", " << "SPV_KHR_shader_atomic_counter_ops" << ", " << "SPV_KHR_shader_ballot" << ", " << "SPV_KHR_shader_clock" << ", " << "SPV_KHR_shader_draw_parameters" << ", " << "SPV_KHR_storage_buffer_storage_class" << ", " << "SPV_KHR_subgroup_vote" << ", " << "SPV_KHR_variable_pointers" << ", " << "SPV_KHR_vulkan_memory_model" << ", " << "SPV_KHR_expect_assume" << ", " << "SPV_KHR_integer_dot_product" << ", " << "SPV_KHR_bit_instructions" << ", " << "SPV_KHR_fragment_shading_rate" << ", " << "SPV_KHR_workgroup_memory_explicit_layout" << ", " << "SPV_KHR_ray_query" << ", " << "SPV_KHR_ray_tracing" << ", " << "SPV_KHR_subgroup_uniform_control_flow" << ", " << "SPV_KHR_linkonce_odr" << ", " << "SPV_KHR_fragment_shader_barycentric" << ", " << "SPV_KHR_ray_cull_mask" << ", " << "SPV_KHR_uniform_group_instructions" << ", " << "SPV_KHR_subgroup_rotate" << ", " << "SPV_EXT_demote_to_helper_invocation" << ", " << "SPV_EXT_descriptor_indexing" << ", " << "SPV_EXT_fragment_fully_covered" << ", " << "SPV_EXT_fragment_invocation_density" << ", " << "SPV_EXT_fragment_shader_interlock" << ", " << "SPV_EXT_physical_storage_buffer" << ", " << "SPV_EXT_shader_stencil_export" << ", " << "SPV_EXT_shader_viewport_index_layer" << ", " << "SPV_EXT_shader_atomic_float_add" << ", " << "SPV_EXT_shader_atomic_float_min_max" << ", " << "SPV_EXT_shader_image_int64" << ", " << "SPV_EXT_shader_atomic_float16_add" << ", " << "SPV_AMD_gpu_shader_half_float_fetch" << ", " << "SPV_AMD_shader_ballot" << ", " << "SPV_AMD_shader_explicit_vertex_parameter" << ", " << "SPV_AMD_shader_fragment_mask" << ", " << "SPV_AMD_shader_image_load_store_lod" << ", " << "SPV_AMD_texture_gather_bias_lod" << ", " << "SPV_AMD_shader_early_and_late_fragment_tests" << ", " << "SPV_GOOGLE_decorate_string" << ", " << "SPV_GOOGLE_hlsl_functionality1" << ", " << "SPV_GOOGLE_user_type" << ", " << "SPV_INTEL_device_side_avc_motion_estimation" << ", " << "SPV_INTEL_media_block_io" << ", " << "SPV_INTEL_shader_integer_functions2" << ", " << "SPV_INTEL_subgroups" << ", " << "SPV_INTEL_vector_compute" << ", " << "SPV_INTEL_float_controls2" << ", " << "SPV_INTEL_function_pointers" << ", " << "SPV_INTEL_inline_assembly" << ", " << "SPV_INTEL_variable_length_array" << ", " << "SPV_INTEL_fpga_memory_attributes" << ", " << "SPV_INTEL_unstructured_loop_controls" << ", " << "SPV_INTEL_fpga_loop_controls" << ", " << "SPV_INTEL_arbitrary_precision_integers" << ", " << "SPV_INTEL_arbitrary_precision_floating_point" << ", " << "SPV_INTEL_kernel_attributes" << ", " << "SPV_INTEL_fpga_memory_accesses" << ", " << "SPV_INTEL_fpga_cluster_attributes" << ", " << "SPV_INTEL_loop_fuse" << ", " << "SPV_INTEL_fpga_buffer_location" << ", " << "SPV_INTEL_arbitrary_precision_fixed_point" << ", " << "SPV_INTEL_usm_storage_classes" << ", " << "SPV_INTEL_io_pipes" << ", " << "SPV_INTEL_blocking_pipes" << ", " << "SPV_INTEL_fpga_reg" << ", " << "SPV_INTEL_long_constant_composite" << ", " << "SPV_INTEL_optnone" << ", " << "SPV_INTEL_debug_module" << ", " << "SPV_INTEL_fp_fast_math_mode" << ", " << "SPV_INTEL_memory_access_aliasing" << ", " << "SPV_INTEL_split_barrier" << ", " << "SPV_INTEL_joint_matrix" << ", " << "SPV_NV_compute_shader_derivatives" << ", " << "SPV_NV_cooperative_matrix" << ", " << "SPV_NV_fragment_shader_barycentric" << ", " << "SPV_NV_geometry_shader_passthrough" << ", " << "SPV_NV_mesh_shader" << ", " << "SPV_NV_ray_tracing" << ", " << "SPV_NV_sample_mask_override_coverage" << ", " << "SPV_NV_shader_image_footprint" << ", " << "SPV_NV_shader_sm_builtins" << ", " << "SPV_NV_shader_subgroup_partitioned" << ", " << "SPV_NV_shading_rate" << ", " << "SPV_NV_stereo_view_rendering" << ", " << "SPV_NV_viewport_array2" << ", " << "SPV_NV_bindless_texture" << ", " << "SPV_NV_ray_tracing_motion_blur" << ", " << "SPV_NVX_multiview_per_view_attributes")};
1521 }();
1522 if (::mlir::failed(_result_value)) {
1523 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ExtensionAttr parameter 'value' which is to be a `::mlir::spirv::Extension`");
1524 return {};
1525 }
1526 // Parse literal '>'
1527 if (odsParser.parseGreater()) return {};
1528 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 1528, __extension__ __PRETTY_FUNCTION__));
1529 return ExtensionAttr::get(odsParser.getContext(),
1530 ::mlir::spirv::Extension((*_result_value)));
1531}
1532
1533void ExtensionAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1534 ::mlir::Builder odsBuilder(getContext());
1535 odsPrinter << "<";
1536 odsPrinter << stringifyExtension(getValue());
1537 odsPrinter << ">";
1538}
1539
1540::mlir::spirv::Extension ExtensionAttr::getValue() const {
1541 return getImpl()->value;
1542}
1543
1544} // namespace spirv
1545} // namespace mlir
1546MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ExtensionAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ExtensionAttr>::id = {}; } }
1547namespace mlir {
1548namespace spirv {
1549namespace detail {
1550struct FunctionControlAttrStorage : public ::mlir::AttributeStorage {
1551 using KeyTy = std::tuple<::mlir::spirv::FunctionControl>;
1552 FunctionControlAttrStorage(::mlir::spirv::FunctionControl value) : value(value) {}
1553
1554 KeyTy getAsKey() const {
1555 return KeyTy(value);
1556 }
1557
1558 bool operator==(const KeyTy &tblgenKey) const {
1559 return (value == std::get<0>(tblgenKey));
1560 }
1561
1562 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1563 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1564 }
1565
1566 static FunctionControlAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1567 auto value = std::get<0>(tblgenKey);
1568 return new (allocator.allocate<FunctionControlAttrStorage>()) FunctionControlAttrStorage(value);
1569 }
1570
1571 ::mlir::spirv::FunctionControl value;
1572};
1573} // namespace detail
1574FunctionControlAttr FunctionControlAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::FunctionControl value) {
1575 return Base::get(context, value);
1576}
1577
1578::mlir::Attribute FunctionControlAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1579 ::mlir::Builder odsBuilder(odsParser.getContext());
1580 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1581 (void) odsLoc;
1582 ::mlir::FailureOr<::mlir::spirv::FunctionControl> _result_value;
1583 // Parse literal '<'
1584 if (odsParser.parseLess()) return {};
1585
1586 // Parse variable 'value'
1587 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::FunctionControl> {
1588 ::mlir::spirv::FunctionControl flags = {};
1589 auto loc = odsParser.getCurrentLocation();
1590 ::llvm::StringRef enumKeyword;
1591 do {
1592 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1593 return ::mlir::failure();
1594 auto maybeEnum = ::mlir::spirv::symbolizeFunctionControl(enumKeyword);
1595 if (!maybeEnum) {
1596 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::FunctionControl" << " to be one of: " << "None" << ", " << "Inline" << ", " << "DontInline" << ", " << "Pure" << ", " << "Const" << ", " << "OptNoneINTEL")};
1597 }
1598 flags = flags | *maybeEnum;
1599 } while(::mlir::succeeded(odsParser.parseOptionalVerticalBar()));
1600 return flags;
1601 }();
1602 if (::mlir::failed(_result_value)) {
1603 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_FunctionControlAttr parameter 'value' which is to be a `::mlir::spirv::FunctionControl`");
1604 return {};
1605 }
1606 // Parse literal '>'
1607 if (odsParser.parseGreater()) return {};
1608 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 1608, __extension__ __PRETTY_FUNCTION__));
1609 return FunctionControlAttr::get(odsParser.getContext(),
1610 ::mlir::spirv::FunctionControl((*_result_value)));
1611}
1612
1613void FunctionControlAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1614 ::mlir::Builder odsBuilder(getContext());
1615 odsPrinter << "<";
1616 odsPrinter << stringifyFunctionControl(getValue());
1617 odsPrinter << ">";
1618}
1619
1620::mlir::spirv::FunctionControl FunctionControlAttr::getValue() const {
1621 return getImpl()->value;
1622}
1623
1624} // namespace spirv
1625} // namespace mlir
1626MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::FunctionControlAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::FunctionControlAttr>::id = {}; } }
1627namespace mlir {
1628namespace spirv {
1629namespace detail {
1630struct GroupOperationAttrStorage : public ::mlir::AttributeStorage {
1631 using KeyTy = std::tuple<::mlir::spirv::GroupOperation>;
1632 GroupOperationAttrStorage(::mlir::spirv::GroupOperation value) : value(value) {}
1633
1634 KeyTy getAsKey() const {
1635 return KeyTy(value);
1636 }
1637
1638 bool operator==(const KeyTy &tblgenKey) const {
1639 return (value == std::get<0>(tblgenKey));
1640 }
1641
1642 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1643 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1644 }
1645
1646 static GroupOperationAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1647 auto value = std::get<0>(tblgenKey);
1648 return new (allocator.allocate<GroupOperationAttrStorage>()) GroupOperationAttrStorage(value);
1649 }
1650
1651 ::mlir::spirv::GroupOperation value;
1652};
1653} // namespace detail
1654GroupOperationAttr GroupOperationAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::GroupOperation value) {
1655 return Base::get(context, value);
1656}
1657
1658::mlir::Attribute GroupOperationAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1659 ::mlir::Builder odsBuilder(odsParser.getContext());
1660 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1661 (void) odsLoc;
1662 ::mlir::FailureOr<::mlir::spirv::GroupOperation> _result_value;
1663 // Parse literal '<'
1664 if (odsParser.parseLess()) return {};
1665
1666 // Parse variable 'value'
1667 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::GroupOperation> {
1668 auto loc = odsParser.getCurrentLocation();
1669 ::llvm::StringRef enumKeyword;
1670 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1671 return ::mlir::failure();
1672 auto maybeEnum = ::mlir::spirv::symbolizeGroupOperation(enumKeyword);
1673 if (maybeEnum)
1674 return *maybeEnum;
1675 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::GroupOperation" << " to be one of: " << "Reduce" << ", " << "InclusiveScan" << ", " << "ExclusiveScan" << ", " << "ClusteredReduce" << ", " << "PartitionedReduceNV" << ", " << "PartitionedInclusiveScanNV" << ", " << "PartitionedExclusiveScanNV")};
1676 }();
1677 if (::mlir::failed(_result_value)) {
1678 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_GroupOperationAttr parameter 'value' which is to be a `::mlir::spirv::GroupOperation`");
1679 return {};
1680 }
1681 // Parse literal '>'
1682 if (odsParser.parseGreater()) return {};
1683 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 1683, __extension__ __PRETTY_FUNCTION__));
1684 return GroupOperationAttr::get(odsParser.getContext(),
1685 ::mlir::spirv::GroupOperation((*_result_value)));
1686}
1687
1688void GroupOperationAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1689 ::mlir::Builder odsBuilder(getContext());
1690 odsPrinter << "<";
1691 odsPrinter << stringifyGroupOperation(getValue());
1692 odsPrinter << ">";
1693}
1694
1695::mlir::spirv::GroupOperation GroupOperationAttr::getValue() const {
1696 return getImpl()->value;
1697}
1698
1699} // namespace spirv
1700} // namespace mlir
1701MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::GroupOperationAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::GroupOperationAttr>::id = {}; } }
1702namespace mlir {
1703namespace spirv {
1704namespace detail {
1705struct ImageFormatAttrStorage : public ::mlir::AttributeStorage {
1706 using KeyTy = std::tuple<::mlir::spirv::ImageFormat>;
1707 ImageFormatAttrStorage(::mlir::spirv::ImageFormat value) : value(value) {}
1708
1709 KeyTy getAsKey() const {
1710 return KeyTy(value);
1711 }
1712
1713 bool operator==(const KeyTy &tblgenKey) const {
1714 return (value == std::get<0>(tblgenKey));
1715 }
1716
1717 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1718 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1719 }
1720
1721 static ImageFormatAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1722 auto value = std::get<0>(tblgenKey);
1723 return new (allocator.allocate<ImageFormatAttrStorage>()) ImageFormatAttrStorage(value);
1724 }
1725
1726 ::mlir::spirv::ImageFormat value;
1727};
1728} // namespace detail
1729ImageFormatAttr ImageFormatAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ImageFormat value) {
1730 return Base::get(context, value);
1731}
1732
1733::mlir::Attribute ImageFormatAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1734 ::mlir::Builder odsBuilder(odsParser.getContext());
1735 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1736 (void) odsLoc;
1737 ::mlir::FailureOr<::mlir::spirv::ImageFormat> _result_value;
1738 // Parse literal '<'
1739 if (odsParser.parseLess()) return {};
1740
1741 // Parse variable 'value'
1742 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ImageFormat> {
1743 auto loc = odsParser.getCurrentLocation();
1744 ::llvm::StringRef enumKeyword;
1745 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1746 return ::mlir::failure();
1747 auto maybeEnum = ::mlir::spirv::symbolizeImageFormat(enumKeyword);
1748 if (maybeEnum)
1749 return *maybeEnum;
1750 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ImageFormat" << " to be one of: " << "Unknown" << ", " << "Rgba32f" << ", " << "Rgba16f" << ", " << "R32f" << ", " << "Rgba8" << ", " << "Rgba8Snorm" << ", " << "Rg32f" << ", " << "Rg16f" << ", " << "R11fG11fB10f" << ", " << "R16f" << ", " << "Rgba16" << ", " << "Rgb10A2" << ", " << "Rg16" << ", " << "Rg8" << ", " << "R16" << ", " << "R8" << ", " << "Rgba16Snorm" << ", " << "Rg16Snorm" << ", " << "Rg8Snorm" << ", " << "R16Snorm" << ", " << "R8Snorm" << ", " << "Rgba32i" << ", " << "Rgba16i" << ", " << "Rgba8i" << ", " << "R32i" << ", " << "Rg32i" << ", " << "Rg16i" << ", " << "Rg8i" << ", " << "R16i" << ", " << "R8i" << ", " << "Rgba32ui" << ", " << "Rgba16ui" << ", " << "Rgba8ui" << ", " << "R32ui" << ", " << "Rgb10a2ui" << ", " << "Rg32ui" << ", " << "Rg16ui" << ", " << "Rg8ui" << ", " << "R16ui" << ", " << "R8ui" << ", " << "R64ui" << ", " << "R64i")};
1751 }();
1752 if (::mlir::failed(_result_value)) {
1753 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ImageFormatAttr parameter 'value' which is to be a `::mlir::spirv::ImageFormat`");
1754 return {};
1755 }
1756 // Parse literal '>'
1757 if (odsParser.parseGreater()) return {};
1758 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 1758, __extension__ __PRETTY_FUNCTION__));
1759 return ImageFormatAttr::get(odsParser.getContext(),
1760 ::mlir::spirv::ImageFormat((*_result_value)));
1761}
1762
1763void ImageFormatAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1764 ::mlir::Builder odsBuilder(getContext());
1765 odsPrinter << "<";
1766 odsPrinter << stringifyImageFormat(getValue());
1767 odsPrinter << ">";
1768}
1769
1770::mlir::spirv::ImageFormat ImageFormatAttr::getValue() const {
1771 return getImpl()->value;
1772}
1773
1774} // namespace spirv
1775} // namespace mlir
1776MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ImageFormatAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ImageFormatAttr>::id = {}; } }
1777namespace mlir {
1778namespace spirv {
1779namespace detail {
1780struct ImageOperandsAttrStorage : public ::mlir::AttributeStorage {
1781 using KeyTy = std::tuple<::mlir::spirv::ImageOperands>;
1782 ImageOperandsAttrStorage(::mlir::spirv::ImageOperands value) : value(value) {}
1783
1784 KeyTy getAsKey() const {
1785 return KeyTy(value);
1786 }
1787
1788 bool operator==(const KeyTy &tblgenKey) const {
1789 return (value == std::get<0>(tblgenKey));
1790 }
1791
1792 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1793 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1794 }
1795
1796 static ImageOperandsAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1797 auto value = std::get<0>(tblgenKey);
1798 return new (allocator.allocate<ImageOperandsAttrStorage>()) ImageOperandsAttrStorage(value);
1799 }
1800
1801 ::mlir::spirv::ImageOperands value;
1802};
1803} // namespace detail
1804ImageOperandsAttr ImageOperandsAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ImageOperands value) {
1805 return Base::get(context, value);
1806}
1807
1808::mlir::Attribute ImageOperandsAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1809 ::mlir::Builder odsBuilder(odsParser.getContext());
1810 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1811 (void) odsLoc;
1812 ::mlir::FailureOr<::mlir::spirv::ImageOperands> _result_value;
1813 // Parse literal '<'
1814 if (odsParser.parseLess()) return {};
1815
1816 // Parse variable 'value'
1817 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ImageOperands> {
1818 ::mlir::spirv::ImageOperands flags = {};
1819 auto loc = odsParser.getCurrentLocation();
1820 ::llvm::StringRef enumKeyword;
1821 do {
1822 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1823 return ::mlir::failure();
1824 auto maybeEnum = ::mlir::spirv::symbolizeImageOperands(enumKeyword);
1825 if (!maybeEnum) {
1826 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ImageOperands" << " to be one of: " << "None" << ", " << "Bias" << ", " << "Lod" << ", " << "Grad" << ", " << "ConstOffset" << ", " << "Offset" << ", " << "ConstOffsets" << ", " << "Sample" << ", " << "MinLod" << ", " << "MakeTexelAvailable" << ", " << "MakeTexelVisible" << ", " << "NonPrivateTexel" << ", " << "VolatileTexel" << ", " << "SignExtend" << ", " << "Offsets" << ", " << "ZeroExtend" << ", " << "Nontemporal")};
1827 }
1828 flags = flags | *maybeEnum;
1829 } while(::mlir::succeeded(odsParser.parseOptionalVerticalBar()));
1830 return flags;
1831 }();
1832 if (::mlir::failed(_result_value)) {
1833 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ImageOperandsAttr parameter 'value' which is to be a `::mlir::spirv::ImageOperands`");
1834 return {};
1835 }
1836 // Parse literal '>'
1837 if (odsParser.parseGreater()) return {};
1838 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 1838, __extension__ __PRETTY_FUNCTION__));
1839 return ImageOperandsAttr::get(odsParser.getContext(),
1840 ::mlir::spirv::ImageOperands((*_result_value)));
1841}
1842
1843void ImageOperandsAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1844 ::mlir::Builder odsBuilder(getContext());
1845 odsPrinter << "<";
1846 odsPrinter << stringifyImageOperands(getValue());
1847 odsPrinter << ">";
1848}
1849
1850::mlir::spirv::ImageOperands ImageOperandsAttr::getValue() const {
1851 return getImpl()->value;
1852}
1853
1854} // namespace spirv
1855} // namespace mlir
1856MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ImageOperandsAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ImageOperandsAttr>::id = {}; } }
1857namespace mlir {
1858namespace spirv {
1859namespace detail {
1860struct JointMatrixPropertiesINTELAttrStorage : public ::mlir::AttributeStorage {
1861 using KeyTy = std::tuple<int, int, int, mlir::Type, mlir::Type, mlir::Type, mlir::Type, mlir::spirv::ScopeAttr>;
1862 JointMatrixPropertiesINTELAttrStorage(int m_size, int n_size, int k_size, mlir::Type a_type, mlir::Type b_type, mlir::Type c_type, mlir::Type result_type, mlir::spirv::ScopeAttr scope) : m_size(m_size), n_size(n_size), k_size(k_size), a_type(a_type), b_type(b_type), c_type(c_type), result_type(result_type), scope(scope) {}
1863
1864 KeyTy getAsKey() const {
1865 return KeyTy(m_size, n_size, k_size, a_type, b_type, c_type, result_type, scope);
1866 }
1867
1868 bool operator==(const KeyTy &tblgenKey) const {
1869 return (m_size == std::get<0>(tblgenKey)) && (n_size == std::get<1>(tblgenKey)) && (k_size == std::get<2>(tblgenKey)) && (a_type == std::get<3>(tblgenKey)) && (b_type == std::get<4>(tblgenKey)) && (c_type == std::get<5>(tblgenKey)) && (result_type == std::get<6>(tblgenKey)) && (scope == std::get<7>(tblgenKey));
1870 }
1871
1872 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1873 return ::llvm::hash_combine(std::get<0>(tblgenKey), std::get<1>(tblgenKey), std::get<2>(tblgenKey), std::get<3>(tblgenKey), std::get<4>(tblgenKey), std::get<5>(tblgenKey), std::get<6>(tblgenKey), std::get<7>(tblgenKey));
1874 }
1875
1876 static JointMatrixPropertiesINTELAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1877 auto m_size = std::get<0>(tblgenKey);
1878 auto n_size = std::get<1>(tblgenKey);
1879 auto k_size = std::get<2>(tblgenKey);
1880 auto a_type = std::get<3>(tblgenKey);
1881 auto b_type = std::get<4>(tblgenKey);
1882 auto c_type = std::get<5>(tblgenKey);
1883 auto result_type = std::get<6>(tblgenKey);
1884 auto scope = std::get<7>(tblgenKey);
1885 return new (allocator.allocate<JointMatrixPropertiesINTELAttrStorage>()) JointMatrixPropertiesINTELAttrStorage(m_size, n_size, k_size, a_type, b_type, c_type, result_type, scope);
1886 }
1887
1888 int m_size;
1889 int n_size;
1890 int k_size;
1891 mlir::Type a_type;
1892 mlir::Type b_type;
1893 mlir::Type c_type;
1894 mlir::Type result_type;
1895 mlir::spirv::ScopeAttr scope;
1896};
1897} // namespace detail
1898JointMatrixPropertiesINTELAttr JointMatrixPropertiesINTELAttr::get(::mlir::MLIRContext *context, int m_size, int n_size, int k_size, mlir::Type a_type, mlir::Type b_type, mlir::Type c_type, mlir::Type result_type, mlir::spirv::ScopeAttr scope) {
1899 return Base::get(context, m_size, n_size, k_size, a_type, b_type, c_type, result_type, scope);
1900}
1901
1902::mlir::Attribute JointMatrixPropertiesINTELAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1903 ::mlir::Builder odsBuilder(odsParser.getContext());
1904 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1905 (void) odsLoc;
1906 ::mlir::FailureOr<int> _result_m_size;
1907 ::mlir::FailureOr<int> _result_n_size;
1908 ::mlir::FailureOr<int> _result_k_size;
1909 ::mlir::FailureOr<mlir::Type> _result_a_type;
1910 ::mlir::FailureOr<mlir::Type> _result_b_type;
1911 ::mlir::FailureOr<mlir::Type> _result_c_type;
1912 ::mlir::FailureOr<mlir::Type> _result_result_type;
1913 ::mlir::FailureOr<mlir::spirv::ScopeAttr> _result_scope;
1914 // Parse literal '<'
1915 if (odsParser.parseLess()) return {};
1916 // Parse parameter struct
1917 bool _seen_m_size = false;
1918 bool _seen_n_size = false;
1919 bool _seen_k_size = false;
1920 bool _seen_a_type = false;
1921 bool _seen_b_type = false;
1922 bool _seen_c_type = false;
1923 bool _seen_result_type = false;
1924 bool _seen_scope = false;
1925 {
1926 const auto _loop_body = [&](::llvm::StringRef _paramKey) -> bool {
1927 // Parse literal '='
1928 if (odsParser.parseEqual()) return {};
1929 if (!_seen_m_size && _paramKey == "m_size") {
1930 _seen_m_size = true;
1931
1932 // Parse variable 'm_size'
1933 _result_m_size = ::mlir::FieldParser<int>::parse(odsParser);
1934 if (::mlir::failed(_result_m_size)) {
1935 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'm_size' which is to be a `int`");
1936 return {};
1937 }
1938 } else if (!_seen_n_size && _paramKey == "n_size") {
1939 _seen_n_size = true;
1940
1941 // Parse variable 'n_size'
1942 _result_n_size = ::mlir::FieldParser<int>::parse(odsParser);
1943 if (::mlir::failed(_result_n_size)) {
1944 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'n_size' which is to be a `int`");
1945 return {};
1946 }
1947 } else if (!_seen_k_size && _paramKey == "k_size") {
1948 _seen_k_size = true;
1949
1950 // Parse variable 'k_size'
1951 _result_k_size = ::mlir::FieldParser<int>::parse(odsParser);
1952 if (::mlir::failed(_result_k_size)) {
1953 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'k_size' which is to be a `int`");
1954 return {};
1955 }
1956 } else if (!_seen_a_type && _paramKey == "a_type") {
1957 _seen_a_type = true;
1958
1959 // Parse variable 'a_type'
1960 _result_a_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
1961 if (::mlir::failed(_result_a_type)) {
1962 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'a_type' which is to be a `mlir::Type`");
1963 return {};
1964 }
1965 } else if (!_seen_b_type && _paramKey == "b_type") {
1966 _seen_b_type = true;
1967
1968 // Parse variable 'b_type'
1969 _result_b_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
1970 if (::mlir::failed(_result_b_type)) {
1971 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'b_type' which is to be a `mlir::Type`");
1972 return {};
1973 }
1974 } else if (!_seen_c_type && _paramKey == "c_type") {
1975 _seen_c_type = true;
1976
1977 // Parse variable 'c_type'
1978 _result_c_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
1979 if (::mlir::failed(_result_c_type)) {
1980 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'c_type' which is to be a `mlir::Type`");
1981 return {};
1982 }
1983 } else if (!_seen_result_type && _paramKey == "result_type") {
1984 _seen_result_type = true;
1985
1986 // Parse variable 'result_type'
1987 _result_result_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
1988 if (::mlir::failed(_result_result_type)) {
1989 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'result_type' which is to be a `mlir::Type`");
1990 return {};
1991 }
1992 } else if (!_seen_scope && _paramKey == "scope") {
1993 _seen_scope = true;
1994
1995 // Parse variable 'scope'
1996 _result_scope = ::mlir::FieldParser<mlir::spirv::ScopeAttr>::parse(odsParser);
1997 if (::mlir::failed(_result_scope)) {
1998 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'scope' which is to be a `mlir::spirv::ScopeAttr`");
1999 return {};
2000 }
2001 } else {
2002 odsParser.emitError(odsParser.getCurrentLocation(), "duplicate or unknown struct parameter name: ") << _paramKey;
2003 return {};
2004 }
2005 return true;
2006 };
2007 for (unsigned odsStructIndex = 0; odsStructIndex < 8; ++odsStructIndex) {
2008 ::llvm::StringRef _paramKey;
2009 if (odsParser.parseKeyword(&_paramKey)) {
2010 odsParser.emitError(odsParser.getCurrentLocation(),
2011 "expected a parameter name in struct");
2012 return {};
2013 }
2014 if (!_loop_body(_paramKey)) return {};
2015 if ((odsStructIndex != 8 - 1) && odsParser.parseComma())
2016 return {};
2017 }
2018 }
2019 // Parse literal '>'
2020 if (odsParser.parseGreater()) return {};
2021 assert(::mlir::succeeded(_result_m_size))(static_cast <bool> (::mlir::succeeded(_result_m_size))
? void (0) : __assert_fail ("::mlir::succeeded(_result_m_size)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2021, __extension__ __PRETTY_FUNCTION__));
2022 assert(::mlir::succeeded(_result_n_size))(static_cast <bool> (::mlir::succeeded(_result_n_size))
? void (0) : __assert_fail ("::mlir::succeeded(_result_n_size)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2022, __extension__ __PRETTY_FUNCTION__));
2023 assert(::mlir::succeeded(_result_k_size))(static_cast <bool> (::mlir::succeeded(_result_k_size))
? void (0) : __assert_fail ("::mlir::succeeded(_result_k_size)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2023, __extension__ __PRETTY_FUNCTION__));
2024 assert(::mlir::succeeded(_result_a_type))(static_cast <bool> (::mlir::succeeded(_result_a_type))
? void (0) : __assert_fail ("::mlir::succeeded(_result_a_type)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2024, __extension__ __PRETTY_FUNCTION__));
2025 assert(::mlir::succeeded(_result_b_type))(static_cast <bool> (::mlir::succeeded(_result_b_type))
? void (0) : __assert_fail ("::mlir::succeeded(_result_b_type)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2025, __extension__ __PRETTY_FUNCTION__));
2026 assert(::mlir::succeeded(_result_c_type))(static_cast <bool> (::mlir::succeeded(_result_c_type))
? void (0) : __assert_fail ("::mlir::succeeded(_result_c_type)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2026, __extension__ __PRETTY_FUNCTION__));
2027 assert(::mlir::succeeded(_result_result_type))(static_cast <bool> (::mlir::succeeded(_result_result_type
)) ? void (0) : __assert_fail ("::mlir::succeeded(_result_result_type)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2027, __extension__ __PRETTY_FUNCTION__));
2028 assert(::mlir::succeeded(_result_scope))(static_cast <bool> (::mlir::succeeded(_result_scope)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_scope)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2028, __extension__ __PRETTY_FUNCTION__));
2029 return JointMatrixPropertiesINTELAttr::get(odsParser.getContext(),
2030 int((*_result_m_size)),
2031 int((*_result_n_size)),
2032 int((*_result_k_size)),
2033 mlir::Type((*_result_a_type)),
2034 mlir::Type((*_result_b_type)),
2035 mlir::Type((*_result_c_type)),
2036 mlir::Type((*_result_result_type)),
2037 mlir::spirv::ScopeAttr((*_result_scope)));
2038}
2039
2040void JointMatrixPropertiesINTELAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2041 ::mlir::Builder odsBuilder(getContext());
2042 odsPrinter << "<";
2043 {
2044 bool _firstPrinted = true;
2045 if (!_firstPrinted) odsPrinter << ", ";
2046 _firstPrinted = false;
2047 odsPrinter << "m_size = ";
2048 odsPrinter.printStrippedAttrOrType(getMSize());
2049 if (!_firstPrinted) odsPrinter << ", ";
2050 _firstPrinted = false;
2051 odsPrinter << "n_size = ";
2052 odsPrinter.printStrippedAttrOrType(getNSize());
2053 if (!_firstPrinted) odsPrinter << ", ";
2054 _firstPrinted = false;
2055 odsPrinter << "k_size = ";
2056 odsPrinter.printStrippedAttrOrType(getKSize());
2057 if (!_firstPrinted) odsPrinter << ", ";
2058 _firstPrinted = false;
2059 odsPrinter << "a_type = ";
2060 odsPrinter.printStrippedAttrOrType(getAType());
2061 if (!_firstPrinted) odsPrinter << ", ";
2062 _firstPrinted = false;
2063 odsPrinter << "b_type = ";
2064 odsPrinter.printStrippedAttrOrType(getBType());
2065 if (!_firstPrinted) odsPrinter << ", ";
2066 _firstPrinted = false;
2067 odsPrinter << "c_type = ";
2068 odsPrinter.printStrippedAttrOrType(getCType());
2069 if (!_firstPrinted) odsPrinter << ", ";
2070 _firstPrinted = false;
2071 odsPrinter << "result_type = ";
2072 odsPrinter.printStrippedAttrOrType(getResultType());
2073 if (!_firstPrinted) odsPrinter << ", ";
2074 _firstPrinted = false;
2075 odsPrinter << "scope = ";
2076 odsPrinter.printStrippedAttrOrType(getScope());
2077 }
2078 odsPrinter << ">";
2079}
2080
2081int JointMatrixPropertiesINTELAttr::getMSize() const {
2082 return getImpl()->m_size;
2083}
2084
2085int JointMatrixPropertiesINTELAttr::getNSize() const {
2086 return getImpl()->n_size;
2087}
2088
2089int JointMatrixPropertiesINTELAttr::getKSize() const {
2090 return getImpl()->k_size;
2091}
2092
2093mlir::Type JointMatrixPropertiesINTELAttr::getAType() const {
2094 return getImpl()->a_type;
2095}
2096
2097mlir::Type JointMatrixPropertiesINTELAttr::getBType() const {
2098 return getImpl()->b_type;
2099}
2100
2101mlir::Type JointMatrixPropertiesINTELAttr::getCType() const {
2102 return getImpl()->c_type;
2103}
2104
2105mlir::Type JointMatrixPropertiesINTELAttr::getResultType() const {
2106 return getImpl()->result_type;
2107}
2108
2109mlir::spirv::ScopeAttr JointMatrixPropertiesINTELAttr::getScope() const {
2110 return getImpl()->scope;
2111}
2112
2113} // namespace spirv
2114} // namespace mlir
2115MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::JointMatrixPropertiesINTELAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::JointMatrixPropertiesINTELAttr>::id = {
}; } }
2116namespace mlir {
2117namespace spirv {
2118namespace detail {
2119struct LinkageTypeAttrStorage : public ::mlir::AttributeStorage {
2120 using KeyTy = std::tuple<::mlir::spirv::LinkageType>;
2121 LinkageTypeAttrStorage(::mlir::spirv::LinkageType value) : value(value) {}
2122
2123 KeyTy getAsKey() const {
2124 return KeyTy(value);
2125 }
2126
2127 bool operator==(const KeyTy &tblgenKey) const {
2128 return (value == std::get<0>(tblgenKey));
2129 }
2130
2131 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2132 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2133 }
2134
2135 static LinkageTypeAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2136 auto value = std::get<0>(tblgenKey);
2137 return new (allocator.allocate<LinkageTypeAttrStorage>()) LinkageTypeAttrStorage(value);
2138 }
2139
2140 ::mlir::spirv::LinkageType value;
2141};
2142} // namespace detail
2143LinkageTypeAttr LinkageTypeAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::LinkageType value) {
2144 return Base::get(context, value);
2145}
2146
2147::mlir::Attribute LinkageTypeAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2148 ::mlir::Builder odsBuilder(odsParser.getContext());
2149 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2150 (void) odsLoc;
2151 ::mlir::FailureOr<::mlir::spirv::LinkageType> _result_value;
2152 // Parse literal '<'
2153 if (odsParser.parseLess()) return {};
2154
2155 // Parse variable 'value'
2156 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::LinkageType> {
2157 auto loc = odsParser.getCurrentLocation();
2158 ::llvm::StringRef enumKeyword;
2159 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2160 return ::mlir::failure();
2161 auto maybeEnum = ::mlir::spirv::symbolizeLinkageType(enumKeyword);
2162 if (maybeEnum)
2163 return *maybeEnum;
2164 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::LinkageType" << " to be one of: " << "Export" << ", " << "Import" << ", " << "LinkOnceODR")};
2165 }();
2166 if (::mlir::failed(_result_value)) {
2167 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_LinkageTypeAttr parameter 'value' which is to be a `::mlir::spirv::LinkageType`");
2168 return {};
2169 }
2170 // Parse literal '>'
2171 if (odsParser.parseGreater()) return {};
2172 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2172, __extension__ __PRETTY_FUNCTION__));
2173 return LinkageTypeAttr::get(odsParser.getContext(),
2174 ::mlir::spirv::LinkageType((*_result_value)));
2175}
2176
2177void LinkageTypeAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2178 ::mlir::Builder odsBuilder(getContext());
2179 odsPrinter << "<";
2180 odsPrinter << stringifyLinkageType(getValue());
2181 odsPrinter << ">";
2182}
2183
2184::mlir::spirv::LinkageType LinkageTypeAttr::getValue() const {
2185 return getImpl()->value;
2186}
2187
2188} // namespace spirv
2189} // namespace mlir
2190MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::LinkageTypeAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::LinkageTypeAttr>::id = {}; } }
2191namespace mlir {
2192namespace spirv {
2193namespace detail {
2194struct LoopControlAttrStorage : public ::mlir::AttributeStorage {
2195 using KeyTy = std::tuple<::mlir::spirv::LoopControl>;
2196 LoopControlAttrStorage(::mlir::spirv::LoopControl value) : value(value) {}
2197
2198 KeyTy getAsKey() const {
2199 return KeyTy(value);
2200 }
2201
2202 bool operator==(const KeyTy &tblgenKey) const {
2203 return (value == std::get<0>(tblgenKey));
2204 }
2205
2206 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2207 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2208 }
2209
2210 static LoopControlAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2211 auto value = std::get<0>(tblgenKey);
2212 return new (allocator.allocate<LoopControlAttrStorage>()) LoopControlAttrStorage(value);
2213 }
2214
2215 ::mlir::spirv::LoopControl value;
2216};
2217} // namespace detail
2218LoopControlAttr LoopControlAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::LoopControl value) {
2219 return Base::get(context, value);
2220}
2221
2222::mlir::Attribute LoopControlAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2223 ::mlir::Builder odsBuilder(odsParser.getContext());
2224 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2225 (void) odsLoc;
2226 ::mlir::FailureOr<::mlir::spirv::LoopControl> _result_value;
2227 // Parse literal '<'
2228 if (odsParser.parseLess()) return {};
2229
2230 // Parse variable 'value'
2231 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::LoopControl> {
2232 ::mlir::spirv::LoopControl flags = {};
2233 auto loc = odsParser.getCurrentLocation();
2234 ::llvm::StringRef enumKeyword;
2235 do {
2236 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2237 return ::mlir::failure();
2238 auto maybeEnum = ::mlir::spirv::symbolizeLoopControl(enumKeyword);
2239 if (!maybeEnum) {
2240 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::LoopControl" << " to be one of: " << "None" << ", " << "Unroll" << ", " << "DontUnroll" << ", " << "DependencyInfinite" << ", " << "DependencyLength" << ", " << "MinIterations" << ", " << "MaxIterations" << ", " << "IterationMultiple" << ", " << "PeelCount" << ", " << "PartialCount" << ", " << "InitiationIntervalINTEL" << ", " << "LoopCoalesceINTEL" << ", " << "MaxConcurrencyINTEL" << ", " << "MaxInterleavingINTEL" << ", " << "DependencyArrayINTEL" << ", " << "SpeculatedIterationsINTEL" << ", " << "PipelineEnableINTEL" << ", " << "NoFusionINTEL")};
2241 }
2242 flags = flags | *maybeEnum;
2243 } while(::mlir::succeeded(odsParser.parseOptionalVerticalBar()));
2244 return flags;
2245 }();
2246 if (::mlir::failed(_result_value)) {
2247 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_LoopControlAttr parameter 'value' which is to be a `::mlir::spirv::LoopControl`");
2248 return {};
2249 }
2250 // Parse literal '>'
2251 if (odsParser.parseGreater()) return {};
2252 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2252, __extension__ __PRETTY_FUNCTION__));
2253 return LoopControlAttr::get(odsParser.getContext(),
2254 ::mlir::spirv::LoopControl((*_result_value)));
2255}
2256
2257void LoopControlAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2258 ::mlir::Builder odsBuilder(getContext());
2259 odsPrinter << "<";
2260 odsPrinter << stringifyLoopControl(getValue());
2261 odsPrinter << ">";
2262}
2263
2264::mlir::spirv::LoopControl LoopControlAttr::getValue() const {
2265 return getImpl()->value;
2266}
2267
2268} // namespace spirv
2269} // namespace mlir
2270MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::LoopControlAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::LoopControlAttr>::id = {}; } }
2271namespace mlir {
2272namespace spirv {
2273namespace detail {
2274struct MatrixLayoutAttrStorage : public ::mlir::AttributeStorage {
2275 using KeyTy = std::tuple<::mlir::spirv::MatrixLayout>;
2276 MatrixLayoutAttrStorage(::mlir::spirv::MatrixLayout value) : value(value) {}
2277
2278 KeyTy getAsKey() const {
2279 return KeyTy(value);
2280 }
2281
2282 bool operator==(const KeyTy &tblgenKey) const {
2283 return (value == std::get<0>(tblgenKey));
2284 }
2285
2286 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2287 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2288 }
2289
2290 static MatrixLayoutAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2291 auto value = std::get<0>(tblgenKey);
2292 return new (allocator.allocate<MatrixLayoutAttrStorage>()) MatrixLayoutAttrStorage(value);
2293 }
2294
2295 ::mlir::spirv::MatrixLayout value;
2296};
2297} // namespace detail
2298MatrixLayoutAttr MatrixLayoutAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::MatrixLayout value) {
2299 return Base::get(context, value);
2300}
2301
2302::mlir::Attribute MatrixLayoutAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2303 ::mlir::Builder odsBuilder(odsParser.getContext());
2304 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2305 (void) odsLoc;
2306 ::mlir::FailureOr<::mlir::spirv::MatrixLayout> _result_value;
2307 // Parse literal '<'
2308 if (odsParser.parseLess()) return {};
2309
2310 // Parse variable 'value'
2311 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::MatrixLayout> {
2312 auto loc = odsParser.getCurrentLocation();
2313 ::llvm::StringRef enumKeyword;
2314 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2315 return ::mlir::failure();
2316 auto maybeEnum = ::mlir::spirv::symbolizeMatrixLayout(enumKeyword);
2317 if (maybeEnum)
2318 return *maybeEnum;
2319 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::MatrixLayout" << " to be one of: " << "ColumnMajor" << ", " << "RowMajor" << ", " << "PackedA" << ", " << "PackedB")};
2320 }();
2321 if (::mlir::failed(_result_value)) {
2322 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_MatrixLayoutAttr parameter 'value' which is to be a `::mlir::spirv::MatrixLayout`");
2323 return {};
2324 }
2325 // Parse literal '>'
2326 if (odsParser.parseGreater()) return {};
2327 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2327, __extension__ __PRETTY_FUNCTION__));
2328 return MatrixLayoutAttr::get(odsParser.getContext(),
2329 ::mlir::spirv::MatrixLayout((*_result_value)));
2330}
2331
2332void MatrixLayoutAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2333 ::mlir::Builder odsBuilder(getContext());
2334 odsPrinter << "<";
2335 odsPrinter << stringifyMatrixLayout(getValue());
2336 odsPrinter << ">";
2337}
2338
2339::mlir::spirv::MatrixLayout MatrixLayoutAttr::getValue() const {
2340 return getImpl()->value;
2341}
2342
2343} // namespace spirv
2344} // namespace mlir
2345MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::MatrixLayoutAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::MatrixLayoutAttr>::id = {}; } }
2346namespace mlir {
2347namespace spirv {
2348namespace detail {
2349struct MemoryAccessAttrStorage : public ::mlir::AttributeStorage {
2350 using KeyTy = std::tuple<::mlir::spirv::MemoryAccess>;
2351 MemoryAccessAttrStorage(::mlir::spirv::MemoryAccess value) : value(value) {}
2352
2353 KeyTy getAsKey() const {
2354 return KeyTy(value);
2355 }
2356
2357 bool operator==(const KeyTy &tblgenKey) const {
2358 return (value == std::get<0>(tblgenKey));
2359 }
2360
2361 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2362 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2363 }
2364
2365 static MemoryAccessAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2366 auto value = std::get<0>(tblgenKey);
2367 return new (allocator.allocate<MemoryAccessAttrStorage>()) MemoryAccessAttrStorage(value);
2368 }
2369
2370 ::mlir::spirv::MemoryAccess value;
2371};
2372} // namespace detail
2373MemoryAccessAttr MemoryAccessAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::MemoryAccess value) {
2374 return Base::get(context, value);
2375}
2376
2377::mlir::Attribute MemoryAccessAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2378 ::mlir::Builder odsBuilder(odsParser.getContext());
2379 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2380 (void) odsLoc;
2381 ::mlir::FailureOr<::mlir::spirv::MemoryAccess> _result_value;
2382 // Parse literal '<'
2383 if (odsParser.parseLess()) return {};
2384
2385 // Parse variable 'value'
2386 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::MemoryAccess> {
2387 ::mlir::spirv::MemoryAccess flags = {};
2388 auto loc = odsParser.getCurrentLocation();
2389 ::llvm::StringRef enumKeyword;
2390 do {
2391 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2392 return ::mlir::failure();
2393 auto maybeEnum = ::mlir::spirv::symbolizeMemoryAccess(enumKeyword);
2394 if (!maybeEnum) {
2395 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::MemoryAccess" << " to be one of: " << "None" << ", " << "Volatile" << ", " << "Aligned" << ", " << "Nontemporal" << ", " << "MakePointerAvailable" << ", " << "MakePointerVisible" << ", " << "NonPrivatePointer" << ", " << "AliasScopeINTELMask" << ", " << "NoAliasINTELMask")};
2396 }
2397 flags = flags | *maybeEnum;
2398 } while(::mlir::succeeded(odsParser.parseOptionalVerticalBar()));
2399 return flags;
2400 }();
2401 if (::mlir::failed(_result_value)) {
2402 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_MemoryAccessAttr parameter 'value' which is to be a `::mlir::spirv::MemoryAccess`");
2403 return {};
2404 }
2405 // Parse literal '>'
2406 if (odsParser.parseGreater()) return {};
2407 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2407, __extension__ __PRETTY_FUNCTION__));
2408 return MemoryAccessAttr::get(odsParser.getContext(),
2409 ::mlir::spirv::MemoryAccess((*_result_value)));
2410}
2411
2412void MemoryAccessAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2413 ::mlir::Builder odsBuilder(getContext());
2414 odsPrinter << "<";
2415 odsPrinter << stringifyMemoryAccess(getValue());
2416 odsPrinter << ">";
2417}
2418
2419::mlir::spirv::MemoryAccess MemoryAccessAttr::getValue() const {
2420 return getImpl()->value;
2421}
2422
2423} // namespace spirv
2424} // namespace mlir
2425MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::MemoryAccessAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::MemoryAccessAttr>::id = {}; } }
2426namespace mlir {
2427namespace spirv {
2428namespace detail {
2429struct MemoryModelAttrStorage : public ::mlir::AttributeStorage {
2430 using KeyTy = std::tuple<::mlir::spirv::MemoryModel>;
2431 MemoryModelAttrStorage(::mlir::spirv::MemoryModel value) : value(value) {}
2432
2433 KeyTy getAsKey() const {
2434 return KeyTy(value);
2435 }
2436
2437 bool operator==(const KeyTy &tblgenKey) const {
2438 return (value == std::get<0>(tblgenKey));
2439 }
2440
2441 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2442 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2443 }
2444
2445 static MemoryModelAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2446 auto value = std::get<0>(tblgenKey);
2447 return new (allocator.allocate<MemoryModelAttrStorage>()) MemoryModelAttrStorage(value);
2448 }
2449
2450 ::mlir::spirv::MemoryModel value;
2451};
2452} // namespace detail
2453MemoryModelAttr MemoryModelAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::MemoryModel value) {
2454 return Base::get(context, value);
2455}
2456
2457::mlir::Attribute MemoryModelAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2458 ::mlir::Builder odsBuilder(odsParser.getContext());
2459 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2460 (void) odsLoc;
2461 ::mlir::FailureOr<::mlir::spirv::MemoryModel> _result_value;
2462 // Parse literal '<'
2463 if (odsParser.parseLess()) return {};
2464
2465 // Parse variable 'value'
2466 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::MemoryModel> {
2467 auto loc = odsParser.getCurrentLocation();
2468 ::llvm::StringRef enumKeyword;
2469 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2470 return ::mlir::failure();
2471 auto maybeEnum = ::mlir::spirv::symbolizeMemoryModel(enumKeyword);
2472 if (maybeEnum)
2473 return *maybeEnum;
2474 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::MemoryModel" << " to be one of: " << "Simple" << ", " << "GLSL450" << ", " << "OpenCL" << ", " << "Vulkan")};
2475 }();
2476 if (::mlir::failed(_result_value)) {
2477 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_MemoryModelAttr parameter 'value' which is to be a `::mlir::spirv::MemoryModel`");
2478 return {};
2479 }
2480 // Parse literal '>'
2481 if (odsParser.parseGreater()) return {};
2482 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2482, __extension__ __PRETTY_FUNCTION__));
2483 return MemoryModelAttr::get(odsParser.getContext(),
2484 ::mlir::spirv::MemoryModel((*_result_value)));
2485}
2486
2487void MemoryModelAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2488 ::mlir::Builder odsBuilder(getContext());
2489 odsPrinter << "<";
2490 odsPrinter << stringifyMemoryModel(getValue());
2491 odsPrinter << ">";
2492}
2493
2494::mlir::spirv::MemoryModel MemoryModelAttr::getValue() const {
2495 return getImpl()->value;
2496}
2497
2498} // namespace spirv
2499} // namespace mlir
2500MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::MemoryModelAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::MemoryModelAttr>::id = {}; } }
2501namespace mlir {
2502namespace spirv {
2503namespace detail {
2504struct MemorySemanticsAttrStorage : public ::mlir::AttributeStorage {
2505 using KeyTy = std::tuple<::mlir::spirv::MemorySemantics>;
2506 MemorySemanticsAttrStorage(::mlir::spirv::MemorySemantics value) : value(value) {}
2507
2508 KeyTy getAsKey() const {
2509 return KeyTy(value);
2510 }
2511
2512 bool operator==(const KeyTy &tblgenKey) const {
2513 return (value == std::get<0>(tblgenKey));
2514 }
2515
2516 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2517 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2518 }
2519
2520 static MemorySemanticsAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2521 auto value = std::get<0>(tblgenKey);
2522 return new (allocator.allocate<MemorySemanticsAttrStorage>()) MemorySemanticsAttrStorage(value);
2523 }
2524
2525 ::mlir::spirv::MemorySemantics value;
2526};
2527} // namespace detail
2528MemorySemanticsAttr MemorySemanticsAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::MemorySemantics value) {
2529 return Base::get(context, value);
2530}
2531
2532::mlir::Attribute MemorySemanticsAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2533 ::mlir::Builder odsBuilder(odsParser.getContext());
2534 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2535 (void) odsLoc;
2536 ::mlir::FailureOr<::mlir::spirv::MemorySemantics> _result_value;
2537 // Parse literal '<'
2538 if (odsParser.parseLess()) return {};
2539
2540 // Parse variable 'value'
2541 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::MemorySemantics> {
2542 ::mlir::spirv::MemorySemantics flags = {};
2543 auto loc = odsParser.getCurrentLocation();
2544 ::llvm::StringRef enumKeyword;
2545 do {
2546 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2547 return ::mlir::failure();
2548 auto maybeEnum = ::mlir::spirv::symbolizeMemorySemantics(enumKeyword);
2549 if (!maybeEnum) {
2550 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::MemorySemantics" << " to be one of: " << "None" << ", " << "Acquire" << ", " << "Release" << ", " << "AcquireRelease" << ", " << "SequentiallyConsistent" << ", " << "UniformMemory" << ", " << "SubgroupMemory" << ", " << "WorkgroupMemory" << ", " << "CrossWorkgroupMemory" << ", " << "AtomicCounterMemory" << ", " << "ImageMemory" << ", " << "OutputMemory" << ", " << "MakeAvailable" << ", " << "MakeVisible" << ", " << "Volatile")};
2551 }
2552 flags = flags | *maybeEnum;
2553 } while(::mlir::succeeded(odsParser.parseOptionalVerticalBar()));
2554 return flags;
2555 }();
2556 if (::mlir::failed(_result_value)) {
2557 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_MemorySemanticsAttr parameter 'value' which is to be a `::mlir::spirv::MemorySemantics`");
2558 return {};
2559 }
2560 // Parse literal '>'
2561 if (odsParser.parseGreater()) return {};
2562 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2562, __extension__ __PRETTY_FUNCTION__));
2563 return MemorySemanticsAttr::get(odsParser.getContext(),
2564 ::mlir::spirv::MemorySemantics((*_result_value)));
2565}
2566
2567void MemorySemanticsAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2568 ::mlir::Builder odsBuilder(getContext());
2569 odsPrinter << "<";
2570 odsPrinter << stringifyMemorySemantics(getValue());
2571 odsPrinter << ">";
2572}
2573
2574::mlir::spirv::MemorySemantics MemorySemanticsAttr::getValue() const {
2575 return getImpl()->value;
2576}
2577
2578} // namespace spirv
2579} // namespace mlir
2580MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::MemorySemanticsAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::MemorySemanticsAttr>::id = {}; } }
2581namespace mlir {
2582namespace spirv {
2583namespace detail {
2584struct OpcodeAttrStorage : public ::mlir::AttributeStorage {
2585 using KeyTy = std::tuple<::mlir::spirv::Opcode>;
2586 OpcodeAttrStorage(::mlir::spirv::Opcode value) : value(value) {}
2587
2588 KeyTy getAsKey() const {
2589 return KeyTy(value);
2590 }
2591
2592 bool operator==(const KeyTy &tblgenKey) const {
2593 return (value == std::get<0>(tblgenKey));
2594 }
2595
2596 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2597 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2598 }
2599
2600 static OpcodeAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2601 auto value = std::get<0>(tblgenKey);
2602 return new (allocator.allocate<OpcodeAttrStorage>()) OpcodeAttrStorage(value);
2603 }
2604
2605 ::mlir::spirv::Opcode value;
2606};
2607} // namespace detail
2608OpcodeAttr OpcodeAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Opcode value) {
2609 return Base::get(context, value);
2610}
2611
2612::mlir::Attribute OpcodeAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2613 ::mlir::Builder odsBuilder(odsParser.getContext());
2614 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2615 (void) odsLoc;
2616 ::mlir::FailureOr<::mlir::spirv::Opcode> _result_value;
2617 // Parse literal '<'
2618 if (odsParser.parseLess()) return {};
2619
2620 // Parse variable 'value'
2621 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Opcode> {
2622 auto loc = odsParser.getCurrentLocation();
2623 ::llvm::StringRef enumKeyword;
2624 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2625 return ::mlir::failure();
2626 auto maybeEnum = ::mlir::spirv::symbolizeOpcode(enumKeyword);
2627 if (maybeEnum)
2628 return *maybeEnum;
2629 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::Opcode" << " to be one of: " << "OpNop" << ", " << "OpUndef" << ", " << "OpSourceContinued" << ", " << "OpSource" << ", " << "OpSourceExtension" << ", " << "OpName" << ", " << "OpMemberName" << ", " << "OpString" << ", " << "OpLine" << ", " << "OpExtension" << ", " << "OpExtInstImport" << ", " << "OpExtInst" << ", " << "OpMemoryModel" << ", " << "OpEntryPoint" << ", " << "OpExecutionMode" << ", " << "OpCapability" << ", " << "OpTypeVoid" << ", " << "OpTypeBool" << ", " << "OpTypeInt" << ", " << "OpTypeFloat" << ", " << "OpTypeVector" << ", " << "OpTypeMatrix" << ", " << "OpTypeImage" << ", " << "OpTypeSampledImage" << ", " << "OpTypeArray" << ", " << "OpTypeRuntimeArray" << ", " << "OpTypeStruct" << ", " << "OpTypePointer" << ", " << "OpTypeFunction" << ", " << "OpTypeForwardPointer" << ", " << "OpConstantTrue" << ", " << "OpConstantFalse" << ", " << "OpConstant" << ", " << "OpConstantComposite" << ", " << "OpConstantNull" << ", " << "OpSpecConstantTrue" << ", " << "OpSpecConstantFalse" << ", " << "OpSpecConstant" << ", " << "OpSpecConstantComposite" << ", " << "OpSpecConstantOp" << ", " << "OpFunction" << ", " << "OpFunctionParameter" << ", " << "OpFunctionEnd" << ", " << "OpFunctionCall" << ", " << "OpVariable" << ", " << "OpLoad" << ", " << "OpStore" << ", " << "OpCopyMemory" << ", " << "OpAccessChain" << ", " << "OpPtrAccessChain" << ", " << "OpInBoundsPtrAccessChain" << ", " << "OpDecorate" << ", " << "OpMemberDecorate" << ", " << "OpVectorExtractDynamic" << ", " << "OpVectorInsertDynamic" << ", " << "OpVectorShuffle" << ", " << "OpCompositeConstruct" << ", " << "OpCompositeExtract" << ", " << "OpCompositeInsert" << ", " << "OpTranspose" << ", " << "OpImageDrefGather" << ", " << "OpImage" << ", " << "OpImageQuerySize" << ", " << "OpConvertFToU" << ", " << "OpConvertFToS" << ", " << "OpConvertSToF" << ", " << "OpConvertUToF" << ", " << "OpUConvert" << ", " << "OpSConvert" << ", " << "OpFConvert" << ", " << "OpPtrCastToGeneric" << ", " << "OpGenericCastToPtr" << ", " << "OpGenericCastToPtrExplicit" << ", " << "OpBitcast" << ", " << "OpSNegate" << ", " << "OpFNegate" << ", " << "OpIAdd" << ", " << "OpFAdd" << ", " << "OpISub" << ", " << "OpFSub" << ", " << "OpIMul" << ", " << "OpFMul" << ", " << "OpUDiv" << ", " << "OpSDiv" << ", " << "OpFDiv" << ", " << "OpUMod" << ", " << "OpSRem" << ", " << "OpSMod" << ", " << "OpFRem" << ", " << "OpFMod" << ", " << "OpVectorTimesScalar" << ", " << "OpMatrixTimesScalar" << ", " << "OpMatrixTimesMatrix" << ", " << "OpIAddCarry" << ", " << "OpISubBorrow" << ", " << "OpIsNan" << ", " << "OpIsInf" << ", " << "OpOrdered" << ", " << "OpUnordered" << ", " << "OpLogicalEqual" << ", " << "OpLogicalNotEqual" << ", " << "OpLogicalOr" << ", " << "OpLogicalAnd" << ", " << "OpLogicalNot" << ", " << "OpSelect" << ", " << "OpIEqual" << ", " << "OpINotEqual" << ", " << "OpUGreaterThan" << ", " << "OpSGreaterThan" << ", " << "OpUGreaterThanEqual" << ", " << "OpSGreaterThanEqual" << ", " << "OpULessThan" << ", " << "OpSLessThan" << ", " << "OpULessThanEqual" << ", " << "OpSLessThanEqual" << ", " << "OpFOrdEqual" << ", " << "OpFUnordEqual" << ", " << "OpFOrdNotEqual" << ", " << "OpFUnordNotEqual" << ", " << "OpFOrdLessThan" << ", " << "OpFUnordLessThan" << ", " << "OpFOrdGreaterThan" << ", " << "OpFUnordGreaterThan" << ", " << "OpFOrdLessThanEqual" << ", " << "OpFUnordLessThanEqual" << ", " << "OpFOrdGreaterThanEqual" << ", " << "OpFUnordGreaterThanEqual" << ", " << "OpShiftRightLogical" << ", " << "OpShiftRightArithmetic" << ", " << "OpShiftLeftLogical" << ", " << "OpBitwiseOr" << ", " << "OpBitwiseXor" << ", " << "OpBitwiseAnd" << ", " << "OpNot" << ", " << "OpBitFieldInsert" << ", " << "OpBitFieldSExtract" << ", " << "OpBitFieldUExtract" << ", " << "OpBitReverse" << ", " << "OpBitCount" << ", " << "OpControlBarrier" << ", " << "OpMemoryBarrier" << ", " << "OpAtomicExchange" << ", " << "OpAtomicCompareExchange" << ", " << "OpAtomicCompareExchangeWeak" << ", " << "OpAtomicIIncrement" << ", " << "OpAtomicIDecrement" << ", " << "OpAtomicIAdd" << ", " << "OpAtomicISub" << ", " << "OpAtomicSMin" << ", " << "OpAtomicUMin" << ", " << "OpAtomicSMax" << ", " << "OpAtomicUMax" << ", " << "OpAtomicAnd" << ", " << "OpAtomicOr" << ", " << "OpAtomicXor" << ", " << "OpPhi" << ", " << "OpLoopMerge" << ", " << "OpSelectionMerge" << ", " << "OpLabel" << ", " << "OpBranch" << ", " << "OpBranchConditional" << ", " << "OpReturn" << ", " << "OpReturnValue" << ", " << "OpUnreachable" << ", " << "OpGroupBroadcast" << ", " << "OpNoLine" << ", " << "OpModuleProcessed" << ", " << "OpGroupNonUniformElect" << ", " << "OpGroupNonUniformBroadcast" << ", " << "OpGroupNonUniformBallot" << ", " << "OpGroupNonUniformShuffle" << ", " << "OpGroupNonUniformShuffleXor" << ", " << "OpGroupNonUniformShuffleUp" << ", " << "OpGroupNonUniformShuffleDown" << ", " << "OpGroupNonUniformIAdd" << ", " << "OpGroupNonUniformFAdd" << ", " << "OpGroupNonUniformIMul" << ", " << "OpGroupNonUniformFMul" << ", " << "OpGroupNonUniformSMin" << ", " << "OpGroupNonUniformUMin" << ", " << "OpGroupNonUniformFMin" << ", " << "OpGroupNonUniformSMax" << ", " << "OpGroupNonUniformUMax" << ", " << "OpGroupNonUniformFMax" << ", " << "OpSubgroupBallotKHR" << ", " << "OpTypeCooperativeMatrixNV" << ", " << "OpCooperativeMatrixLoadNV" << ", " << "OpCooperativeMatrixStoreNV" << ", " << "OpCooperativeMatrixMulAddNV" << ", " << "OpCooperativeMatrixLengthNV" << ", " << "OpSubgroupBlockReadINTEL" << ", " << "OpSubgroupBlockWriteINTEL" << ", " << "OpAssumeTrueKHR" << ", " << "OpAtomicFAddEXT" << ", " << "OpTypeJointMatrixINTEL" << ", " << "OpJointMatrixLoadINTEL" << ", " << "OpJointMatrixStoreINTEL" << ", " << "OpJointMatrixMadINTEL" << ", " << "OpJointMatrixWorkItemLengthINTEL")};
2630 }();
2631 if (::mlir::failed(_result_value)) {
2632 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_OpcodeAttr parameter 'value' which is to be a `::mlir::spirv::Opcode`");
2633 return {};
2634 }
2635 // Parse literal '>'
2636 if (odsParser.parseGreater()) return {};
2637 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2637, __extension__ __PRETTY_FUNCTION__));
2638 return OpcodeAttr::get(odsParser.getContext(),
2639 ::mlir::spirv::Opcode((*_result_value)));
2640}
2641
2642void OpcodeAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2643 ::mlir::Builder odsBuilder(getContext());
2644 odsPrinter << "<";
2645 odsPrinter << stringifyOpcode(getValue());
2646 odsPrinter << ">";
2647}
2648
2649::mlir::spirv::Opcode OpcodeAttr::getValue() const {
2650 return getImpl()->value;
2651}
2652
2653} // namespace spirv
2654} // namespace mlir
2655MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::OpcodeAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::OpcodeAttr>::id = {}; } }
2656namespace mlir {
2657namespace spirv {
2658namespace detail {
2659struct ResourceLimitsAttrStorage : public ::mlir::AttributeStorage {
2660 using KeyTy = std::tuple<int, int, ArrayAttr, int, ArrayAttr>;
2661 ResourceLimitsAttrStorage(int max_compute_shared_memory_size, int max_compute_workgroup_invocations, ArrayAttr max_compute_workgroup_size, int subgroup_size, ArrayAttr cooperative_matrix_properties_nv) : max_compute_shared_memory_size(max_compute_shared_memory_size), max_compute_workgroup_invocations(max_compute_workgroup_invocations), max_compute_workgroup_size(max_compute_workgroup_size), subgroup_size(subgroup_size), cooperative_matrix_properties_nv(cooperative_matrix_properties_nv) {}
2662
2663 KeyTy getAsKey() const {
2664 return KeyTy(max_compute_shared_memory_size, max_compute_workgroup_invocations, max_compute_workgroup_size, subgroup_size, cooperative_matrix_properties_nv);
2665 }
2666
2667 bool operator==(const KeyTy &tblgenKey) const {
2668 return (max_compute_shared_memory_size == std::get<0>(tblgenKey)) && (max_compute_workgroup_invocations == std::get<1>(tblgenKey)) && (max_compute_workgroup_size == std::get<2>(tblgenKey)) && (subgroup_size == std::get<3>(tblgenKey)) && (cooperative_matrix_properties_nv == std::get<4>(tblgenKey));
2669 }
2670
2671 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2672 return ::llvm::hash_combine(std::get<0>(tblgenKey), std::get<1>(tblgenKey), std::get<2>(tblgenKey), std::get<3>(tblgenKey), std::get<4>(tblgenKey));
2673 }
2674
2675 static ResourceLimitsAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2676 auto max_compute_shared_memory_size = std::get<0>(tblgenKey);
2677 auto max_compute_workgroup_invocations = std::get<1>(tblgenKey);
2678 auto max_compute_workgroup_size = std::get<2>(tblgenKey);
2679 auto subgroup_size = std::get<3>(tblgenKey);
2680 auto cooperative_matrix_properties_nv = std::get<4>(tblgenKey);
2681 return new (allocator.allocate<ResourceLimitsAttrStorage>()) ResourceLimitsAttrStorage(max_compute_shared_memory_size, max_compute_workgroup_invocations, max_compute_workgroup_size, subgroup_size, cooperative_matrix_properties_nv);
2682 }
2683
2684 int max_compute_shared_memory_size;
2685 int max_compute_workgroup_invocations;
2686 ArrayAttr max_compute_workgroup_size;
2687 int subgroup_size;
2688 ArrayAttr cooperative_matrix_properties_nv;
2689};
2690} // namespace detail
2691ResourceLimitsAttr ResourceLimitsAttr::get(::mlir::MLIRContext *context, int max_compute_shared_memory_size, int max_compute_workgroup_invocations, ArrayAttr max_compute_workgroup_size, int subgroup_size, ArrayAttr cooperative_matrix_properties_nv) {
2692 return Base::get(context, max_compute_shared_memory_size, max_compute_workgroup_invocations, max_compute_workgroup_size, subgroup_size, cooperative_matrix_properties_nv);
2693}
2694
2695::mlir::Attribute ResourceLimitsAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2696 ::mlir::Builder odsBuilder(odsParser.getContext());
2697 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2698 (void) odsLoc;
2699 ::mlir::FailureOr<int> _result_max_compute_shared_memory_size;
2700 ::mlir::FailureOr<int> _result_max_compute_workgroup_invocations;
2701 ::mlir::FailureOr<ArrayAttr> _result_max_compute_workgroup_size;
2702 ::mlir::FailureOr<int> _result_subgroup_size;
2703 ::mlir::FailureOr<ArrayAttr> _result_cooperative_matrix_properties_nv;
2704 // Parse literal '<'
2705 if (odsParser.parseLess()) return {};
2706 // Parse parameter struct
2707 bool _seen_max_compute_shared_memory_size = false;
2708 bool _seen_max_compute_workgroup_invocations = false;
2709 bool _seen_max_compute_workgroup_size = false;
2710 bool _seen_subgroup_size = false;
2711 bool _seen_cooperative_matrix_properties_nv = false;
2712 {
2713 const auto _loop_body = [&](::llvm::StringRef _paramKey) -> bool {
2714 // Parse literal '='
2715 if (odsParser.parseEqual()) return {};
2716 if (!_seen_max_compute_shared_memory_size && _paramKey == "max_compute_shared_memory_size") {
2717 _seen_max_compute_shared_memory_size = true;
2718
2719 // Parse variable 'max_compute_shared_memory_size'
2720 _result_max_compute_shared_memory_size = ::mlir::FieldParser<int>::parse(odsParser);
2721 if (::mlir::failed(_result_max_compute_shared_memory_size)) {
2722 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ResourceLimitsAttr parameter 'max_compute_shared_memory_size' which is to be a `int`");
2723 return {};
2724 }
2725 } else if (!_seen_max_compute_workgroup_invocations && _paramKey == "max_compute_workgroup_invocations") {
2726 _seen_max_compute_workgroup_invocations = true;
2727
2728 // Parse variable 'max_compute_workgroup_invocations'
2729 _result_max_compute_workgroup_invocations = ::mlir::FieldParser<int>::parse(odsParser);
2730 if (::mlir::failed(_result_max_compute_workgroup_invocations)) {
2731 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ResourceLimitsAttr parameter 'max_compute_workgroup_invocations' which is to be a `int`");
2732 return {};
2733 }
2734 } else if (!_seen_max_compute_workgroup_size && _paramKey == "max_compute_workgroup_size") {
2735 _seen_max_compute_workgroup_size = true;
2736
2737 // Parse variable 'max_compute_workgroup_size'
2738 _result_max_compute_workgroup_size = ::mlir::FieldParser<ArrayAttr>::parse(odsParser);
2739 if (::mlir::failed(_result_max_compute_workgroup_size)) {
2740 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ResourceLimitsAttr parameter 'max_compute_workgroup_size' which is to be a `ArrayAttr`");
2741 return {};
2742 }
2743 } else if (!_seen_subgroup_size && _paramKey == "subgroup_size") {
2744 _seen_subgroup_size = true;
2745
2746 // Parse variable 'subgroup_size'
2747 _result_subgroup_size = ::mlir::FieldParser<int>::parse(odsParser);
2748 if (::mlir::failed(_result_subgroup_size)) {
2749 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ResourceLimitsAttr parameter 'subgroup_size' which is to be a `int`");
2750 return {};
2751 }
2752 } else if (!_seen_cooperative_matrix_properties_nv && _paramKey == "cooperative_matrix_properties_nv") {
2753 _seen_cooperative_matrix_properties_nv = true;
2754
2755 // Parse variable 'cooperative_matrix_properties_nv'
2756 _result_cooperative_matrix_properties_nv = ::mlir::FieldParser<ArrayAttr>::parse(odsParser);
2757 if (::mlir::failed(_result_cooperative_matrix_properties_nv)) {
2758 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ResourceLimitsAttr parameter 'cooperative_matrix_properties_nv' which is to be a `ArrayAttr`");
2759 return {};
2760 }
2761 } else {
2762 odsParser.emitError(odsParser.getCurrentLocation(), "duplicate or unknown struct parameter name: ") << _paramKey;
2763 return {};
2764 }
2765 return true;
2766 };
2767 ::llvm::StringRef _paramKey;
2768 if (!odsParser.parseOptionalKeyword(&_paramKey)) {
2769 if (!_loop_body(_paramKey)) return {};
2770 while (!odsParser.parseOptionalComma()) {
2771 ::llvm::StringRef _paramKey;
2772 if (odsParser.parseKeyword(&_paramKey)) {
2773 odsParser.emitError(odsParser.getCurrentLocation(),
2774 "expected a parameter name in struct");
2775 return {};
2776 }
2777 if (!_loop_body(_paramKey)) return {};
2778 }
2779 }
2780 }
2781 // Parse literal '>'
2782 if (odsParser.parseGreater()) return {};
2783 return ResourceLimitsAttr::get(odsParser.getContext(),
2784 int((_result_max_compute_shared_memory_size.value_or(16384))),
2785 int((_result_max_compute_workgroup_invocations.value_or(128))),
2786 ArrayAttr((_result_max_compute_workgroup_size.value_or(odsBuilder.getI32ArrayAttr({128, 128, 64})))),
2787 int((_result_subgroup_size.value_or(32))),
2788 ArrayAttr((_result_cooperative_matrix_properties_nv.value_or(nullptr))));
2789}
2790
2791void ResourceLimitsAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2792 ::mlir::Builder odsBuilder(getContext());
2793 odsPrinter << "<";
2794 {
2795 bool _firstPrinted = true;
2796 if (!(getMaxComputeSharedMemorySize() == 16384)) {
2797 if (!_firstPrinted) odsPrinter << ", ";
2798 _firstPrinted = false;
2799 odsPrinter << "max_compute_shared_memory_size = ";
2800 if (!(getMaxComputeSharedMemorySize() == 16384)) {
2801 odsPrinter.printStrippedAttrOrType(getMaxComputeSharedMemorySize());
2802 }
2803 }
2804 if (!(getMaxComputeWorkgroupInvocations() == 128)) {
2805 if (!_firstPrinted) odsPrinter << ", ";
2806 _firstPrinted = false;
2807 odsPrinter << "max_compute_workgroup_invocations = ";
2808 if (!(getMaxComputeWorkgroupInvocations() == 128)) {
2809 odsPrinter.printStrippedAttrOrType(getMaxComputeWorkgroupInvocations());
2810 }
2811 }
2812 if (!(getMaxComputeWorkgroupSize() == odsBuilder.getI32ArrayAttr({128, 128, 64}))) {
2813 if (!_firstPrinted) odsPrinter << ", ";
2814 _firstPrinted = false;
2815 odsPrinter << "max_compute_workgroup_size = ";
2816 if (!(getMaxComputeWorkgroupSize() == odsBuilder.getI32ArrayAttr({128, 128, 64}))) {
2817 odsPrinter.printStrippedAttrOrType(getMaxComputeWorkgroupSize());
2818 }
2819 }
2820 if (!(getSubgroupSize() == 32)) {
2821 if (!_firstPrinted) odsPrinter << ", ";
2822 _firstPrinted = false;
2823 odsPrinter << "subgroup_size = ";
2824 if (!(getSubgroupSize() == 32)) {
2825 odsPrinter.printStrippedAttrOrType(getSubgroupSize());
2826 }
2827 }
2828 if (!(getCooperativeMatrixPropertiesNv() == nullptr)) {
2829 if (!_firstPrinted) odsPrinter << ", ";
2830 _firstPrinted = false;
Value stored to '_firstPrinted' is never read
2831 odsPrinter << "cooperative_matrix_properties_nv = ";
2832 if (!(getCooperativeMatrixPropertiesNv() == nullptr)) {
2833 odsPrinter.printStrippedAttrOrType(getCooperativeMatrixPropertiesNv());
2834 }
2835 }
2836 }
2837 odsPrinter << ">";
2838}
2839
2840int ResourceLimitsAttr::getMaxComputeSharedMemorySize() const {
2841 return getImpl()->max_compute_shared_memory_size;
2842}
2843
2844int ResourceLimitsAttr::getMaxComputeWorkgroupInvocations() const {
2845 return getImpl()->max_compute_workgroup_invocations;
2846}
2847
2848ArrayAttr ResourceLimitsAttr::getMaxComputeWorkgroupSize() const {
2849 return getImpl()->max_compute_workgroup_size;
2850}
2851
2852int ResourceLimitsAttr::getSubgroupSize() const {
2853 return getImpl()->subgroup_size;
2854}
2855
2856ArrayAttr ResourceLimitsAttr::getCooperativeMatrixPropertiesNv() const {
2857 return getImpl()->cooperative_matrix_properties_nv;
2858}
2859
2860} // namespace spirv
2861} // namespace mlir
2862MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ResourceLimitsAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ResourceLimitsAttr>::id = {}; } }
2863namespace mlir {
2864namespace spirv {
2865namespace detail {
2866struct ImageSamplerUseInfoAttrStorage : public ::mlir::AttributeStorage {
2867 using KeyTy = std::tuple<::mlir::spirv::ImageSamplerUseInfo>;
2868 ImageSamplerUseInfoAttrStorage(::mlir::spirv::ImageSamplerUseInfo value) : value(value) {}
2869
2870 KeyTy getAsKey() const {
2871 return KeyTy(value);
2872 }
2873
2874 bool operator==(const KeyTy &tblgenKey) const {
2875 return (value == std::get<0>(tblgenKey));
2876 }
2877
2878 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2879 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2880 }
2881
2882 static ImageSamplerUseInfoAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2883 auto value = std::get<0>(tblgenKey);
2884 return new (allocator.allocate<ImageSamplerUseInfoAttrStorage>()) ImageSamplerUseInfoAttrStorage(value);
2885 }
2886
2887 ::mlir::spirv::ImageSamplerUseInfo value;
2888};
2889} // namespace detail
2890ImageSamplerUseInfoAttr ImageSamplerUseInfoAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ImageSamplerUseInfo value) {
2891 return Base::get(context, value);
2892}
2893
2894::mlir::Attribute ImageSamplerUseInfoAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2895 ::mlir::Builder odsBuilder(odsParser.getContext());
2896 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2897 (void) odsLoc;
2898 ::mlir::FailureOr<::mlir::spirv::ImageSamplerUseInfo> _result_value;
2899 // Parse literal '<'
2900 if (odsParser.parseLess()) return {};
2901
2902 // Parse variable 'value'
2903 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ImageSamplerUseInfo> {
2904 auto loc = odsParser.getCurrentLocation();
2905 ::llvm::StringRef enumKeyword;
2906 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2907 return ::mlir::failure();
2908 auto maybeEnum = ::mlir::spirv::symbolizeImageSamplerUseInfo(enumKeyword);
2909 if (maybeEnum)
2910 return *maybeEnum;
2911 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ImageSamplerUseInfo" << " to be one of: " << "SamplerUnknown" << ", " << "NeedSampler" << ", " << "NoSampler")};
2912 }();
2913 if (::mlir::failed(_result_value)) {
2914 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_SamplerUseAttr parameter 'value' which is to be a `::mlir::spirv::ImageSamplerUseInfo`");
2915 return {};
2916 }
2917 // Parse literal '>'
2918 if (odsParser.parseGreater()) return {};
2919 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2919, __extension__ __PRETTY_FUNCTION__));
2920 return ImageSamplerUseInfoAttr::get(odsParser.getContext(),
2921 ::mlir::spirv::ImageSamplerUseInfo((*_result_value)));
2922}
2923
2924void ImageSamplerUseInfoAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2925 ::mlir::Builder odsBuilder(getContext());
2926 odsPrinter << "<";
2927 odsPrinter << stringifyImageSamplerUseInfo(getValue());
2928 odsPrinter << ">";
2929}
2930
2931::mlir::spirv::ImageSamplerUseInfo ImageSamplerUseInfoAttr::getValue() const {
2932 return getImpl()->value;
2933}
2934
2935} // namespace spirv
2936} // namespace mlir
2937MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ImageSamplerUseInfoAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ImageSamplerUseInfoAttr>::id = {}; } }
2938namespace mlir {
2939namespace spirv {
2940namespace detail {
2941struct ImageSamplingInfoAttrStorage : public ::mlir::AttributeStorage {
2942 using KeyTy = std::tuple<::mlir::spirv::ImageSamplingInfo>;
2943 ImageSamplingInfoAttrStorage(::mlir::spirv::ImageSamplingInfo value) : value(value) {}
2944
2945 KeyTy getAsKey() const {
2946 return KeyTy(value);
2947 }
2948
2949 bool operator==(const KeyTy &tblgenKey) const {
2950 return (value == std::get<0>(tblgenKey));
2951 }
2952
2953 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2954 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2955 }
2956
2957 static ImageSamplingInfoAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2958 auto value = std::get<0>(tblgenKey);
2959 return new (allocator.allocate<ImageSamplingInfoAttrStorage>()) ImageSamplingInfoAttrStorage(value);
2960 }
2961
2962 ::mlir::spirv::ImageSamplingInfo value;
2963};
2964} // namespace detail
2965ImageSamplingInfoAttr ImageSamplingInfoAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ImageSamplingInfo value) {
2966 return Base::get(context, value);
2967}
2968
2969::mlir::Attribute ImageSamplingInfoAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2970 ::mlir::Builder odsBuilder(odsParser.getContext());
2971 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2972 (void) odsLoc;
2973 ::mlir::FailureOr<::mlir::spirv::ImageSamplingInfo> _result_value;
2974 // Parse literal '<'
2975 if (odsParser.parseLess()) return {};
2976
2977 // Parse variable 'value'
2978 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ImageSamplingInfo> {
2979 auto loc = odsParser.getCurrentLocation();
2980 ::llvm::StringRef enumKeyword;
2981 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2982 return ::mlir::failure();
2983 auto maybeEnum = ::mlir::spirv::symbolizeImageSamplingInfo(enumKeyword);
2984 if (maybeEnum)
2985 return *maybeEnum;
2986 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ImageSamplingInfo" << " to be one of: " << "SingleSampled" << ", " << "MultiSampled")};
2987 }();
2988 if (::mlir::failed(_result_value)) {
2989 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_SamplingAttr parameter 'value' which is to be a `::mlir::spirv::ImageSamplingInfo`");
2990 return {};
2991 }
2992 // Parse literal '>'
2993 if (odsParser.parseGreater()) return {};
2994 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 2994, __extension__ __PRETTY_FUNCTION__));
2995 return ImageSamplingInfoAttr::get(odsParser.getContext(),
2996 ::mlir::spirv::ImageSamplingInfo((*_result_value)));
2997}
2998
2999void ImageSamplingInfoAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3000 ::mlir::Builder odsBuilder(getContext());
3001 odsPrinter << "<";
3002 odsPrinter << stringifyImageSamplingInfo(getValue());
3003 odsPrinter << ">";
3004}
3005
3006::mlir::spirv::ImageSamplingInfo ImageSamplingInfoAttr::getValue() const {
3007 return getImpl()->value;
3008}
3009
3010} // namespace spirv
3011} // namespace mlir
3012MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ImageSamplingInfoAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ImageSamplingInfoAttr>::id = {}; } }
3013namespace mlir {
3014namespace spirv {
3015namespace detail {
3016struct ScopeAttrStorage : public ::mlir::AttributeStorage {
3017 using KeyTy = std::tuple<::mlir::spirv::Scope>;
3018 ScopeAttrStorage(::mlir::spirv::Scope value) : value(value) {}
3019
3020 KeyTy getAsKey() const {
3021 return KeyTy(value);
3022 }
3023
3024 bool operator==(const KeyTy &tblgenKey) const {
3025 return (value == std::get<0>(tblgenKey));
3026 }
3027
3028 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
3029 return ::llvm::hash_combine(std::get<0>(tblgenKey));
3030 }
3031
3032 static ScopeAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
3033 auto value = std::get<0>(tblgenKey);
3034 return new (allocator.allocate<ScopeAttrStorage>()) ScopeAttrStorage(value);
3035 }
3036
3037 ::mlir::spirv::Scope value;
3038};
3039} // namespace detail
3040ScopeAttr ScopeAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Scope value) {
3041 return Base::get(context, value);
3042}
3043
3044::mlir::Attribute ScopeAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
3045 ::mlir::Builder odsBuilder(odsParser.getContext());
3046 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
3047 (void) odsLoc;
3048 ::mlir::FailureOr<::mlir::spirv::Scope> _result_value;
3049 // Parse literal '<'
3050 if (odsParser.parseLess()) return {};
3051
3052 // Parse variable 'value'
3053 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Scope> {
3054 auto loc = odsParser.getCurrentLocation();
3055 ::llvm::StringRef enumKeyword;
3056 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
3057 return ::mlir::failure();
3058 auto maybeEnum = ::mlir::spirv::symbolizeScope(enumKeyword);
3059 if (maybeEnum)
3060 return *maybeEnum;
3061 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::Scope" << " to be one of: " << "CrossDevice" << ", " << "Device" << ", " << "Workgroup" << ", " << "Subgroup" << ", " << "Invocation" << ", " << "QueueFamily" << ", " << "ShaderCallKHR")};
3062 }();
3063 if (::mlir::failed(_result_value)) {
3064 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ScopeAttr parameter 'value' which is to be a `::mlir::spirv::Scope`");
3065 return {};
3066 }
3067 // Parse literal '>'
3068 if (odsParser.parseGreater()) return {};
3069 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 3069, __extension__ __PRETTY_FUNCTION__));
3070 return ScopeAttr::get(odsParser.getContext(),
3071 ::mlir::spirv::Scope((*_result_value)));
3072}
3073
3074void ScopeAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3075 ::mlir::Builder odsBuilder(getContext());
3076 odsPrinter << "<";
3077 odsPrinter << stringifyScope(getValue());
3078 odsPrinter << ">";
3079}
3080
3081::mlir::spirv::Scope ScopeAttr::getValue() const {
3082 return getImpl()->value;
3083}
3084
3085} // namespace spirv
3086} // namespace mlir
3087MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ScopeAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ScopeAttr>::id = {}; } }
3088namespace mlir {
3089namespace spirv {
3090namespace detail {
3091struct SelectionControlAttrStorage : public ::mlir::AttributeStorage {
3092 using KeyTy = std::tuple<::mlir::spirv::SelectionControl>;
3093 SelectionControlAttrStorage(::mlir::spirv::SelectionControl value) : value(value) {}
3094
3095 KeyTy getAsKey() const {
3096 return KeyTy(value);
3097 }
3098
3099 bool operator==(const KeyTy &tblgenKey) const {
3100 return (value == std::get<0>(tblgenKey));
3101 }
3102
3103 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
3104 return ::llvm::hash_combine(std::get<0>(tblgenKey));
3105 }
3106
3107 static SelectionControlAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
3108 auto value = std::get<0>(tblgenKey);
3109 return new (allocator.allocate<SelectionControlAttrStorage>()) SelectionControlAttrStorage(value);
3110 }
3111
3112 ::mlir::spirv::SelectionControl value;
3113};
3114} // namespace detail
3115SelectionControlAttr SelectionControlAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::SelectionControl value) {
3116 return Base::get(context, value);
3117}
3118
3119::mlir::Attribute SelectionControlAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
3120 ::mlir::Builder odsBuilder(odsParser.getContext());
3121 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
3122 (void) odsLoc;
3123 ::mlir::FailureOr<::mlir::spirv::SelectionControl> _result_value;
3124 // Parse literal '<'
3125 if (odsParser.parseLess()) return {};
3126
3127 // Parse variable 'value'
3128 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::SelectionControl> {
3129 ::mlir::spirv::SelectionControl flags = {};
3130 auto loc = odsParser.getCurrentLocation();
3131 ::llvm::StringRef enumKeyword;
3132 do {
3133 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
3134 return ::mlir::failure();
3135 auto maybeEnum = ::mlir::spirv::symbolizeSelectionControl(enumKeyword);
3136 if (!maybeEnum) {
3137 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::SelectionControl" << " to be one of: " << "None" << ", " << "Flatten" << ", " << "DontFlatten")};
3138 }
3139 flags = flags | *maybeEnum;
3140 } while(::mlir::succeeded(odsParser.parseOptionalVerticalBar()));
3141 return flags;
3142 }();
3143 if (::mlir::failed(_result_value)) {
3144 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_SelectionControlAttr parameter 'value' which is to be a `::mlir::spirv::SelectionControl`");
3145 return {};
3146 }
3147 // Parse literal '>'
3148 if (odsParser.parseGreater()) return {};
3149 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 3149, __extension__ __PRETTY_FUNCTION__));
3150 return SelectionControlAttr::get(odsParser.getContext(),
3151 ::mlir::spirv::SelectionControl((*_result_value)));
3152}
3153
3154void SelectionControlAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3155 ::mlir::Builder odsBuilder(getContext());
3156 odsPrinter << "<";
3157 odsPrinter << stringifySelectionControl(getValue());
3158 odsPrinter << ">";
3159}
3160
3161::mlir::spirv::SelectionControl SelectionControlAttr::getValue() const {
3162 return getImpl()->value;
3163}
3164
3165} // namespace spirv
3166} // namespace mlir
3167MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::SelectionControlAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::SelectionControlAttr>::id = {}; } }
3168namespace mlir {
3169namespace spirv {
3170namespace detail {
3171struct StorageClassAttrStorage : public ::mlir::AttributeStorage {
3172 using KeyTy = std::tuple<::mlir::spirv::StorageClass>;
3173 StorageClassAttrStorage(::mlir::spirv::StorageClass value) : value(value) {}
3174
3175 KeyTy getAsKey() const {
3176 return KeyTy(value);
3177 }
3178
3179 bool operator==(const KeyTy &tblgenKey) const {
3180 return (value == std::get<0>(tblgenKey));
3181 }
3182
3183 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
3184 return ::llvm::hash_combine(std::get<0>(tblgenKey));
3185 }
3186
3187 static StorageClassAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
3188 auto value = std::get<0>(tblgenKey);
3189 return new (allocator.allocate<StorageClassAttrStorage>()) StorageClassAttrStorage(value);
3190 }
3191
3192 ::mlir::spirv::StorageClass value;
3193};
3194} // namespace detail
3195StorageClassAttr StorageClassAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::StorageClass value) {
3196 return Base::get(context, value);
3197}
3198
3199::mlir::Attribute StorageClassAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
3200 ::mlir::Builder odsBuilder(odsParser.getContext());
3201 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
3202 (void) odsLoc;
3203 ::mlir::FailureOr<::mlir::spirv::StorageClass> _result_value;
3204 // Parse literal '<'
3205 if (odsParser.parseLess()) return {};
3206
3207 // Parse variable 'value'
3208 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::StorageClass> {
3209 auto loc = odsParser.getCurrentLocation();
3210 ::llvm::StringRef enumKeyword;
3211 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
3212 return ::mlir::failure();
3213 auto maybeEnum = ::mlir::spirv::symbolizeStorageClass(enumKeyword);
3214 if (maybeEnum)
3215 return *maybeEnum;
3216 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::StorageClass" << " to be one of: " << "UniformConstant" << ", " << "Input" << ", " << "Uniform" << ", " << "Output" << ", " << "Workgroup" << ", " << "CrossWorkgroup" << ", " << "Private" << ", " << "Function" << ", " << "Generic" << ", " << "PushConstant" << ", " << "AtomicCounter" << ", " << "Image" << ", " << "StorageBuffer" << ", " << "CallableDataKHR" << ", " << "IncomingCallableDataKHR" << ", " << "RayPayloadKHR" << ", " << "HitAttributeKHR" << ", " << "IncomingRayPayloadKHR" << ", " << "ShaderRecordBufferKHR" << ", " << "PhysicalStorageBuffer" << ", " << "CodeSectionINTEL" << ", " << "DeviceOnlyINTEL" << ", " << "HostOnlyINTEL")};
3217 }();
3218 if (::mlir::failed(_result_value)) {
3219 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_StorageClassAttr parameter 'value' which is to be a `::mlir::spirv::StorageClass`");
3220 return {};
3221 }
3222 // Parse literal '>'
3223 if (odsParser.parseGreater()) return {};
3224 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 3224, __extension__ __PRETTY_FUNCTION__));
3225 return StorageClassAttr::get(odsParser.getContext(),
3226 ::mlir::spirv::StorageClass((*_result_value)));
3227}
3228
3229void StorageClassAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3230 ::mlir::Builder odsBuilder(getContext());
3231 odsPrinter << "<";
3232 odsPrinter << stringifyStorageClass(getValue());
3233 odsPrinter << ">";
3234}
3235
3236::mlir::spirv::StorageClass StorageClassAttr::getValue() const {
3237 return getImpl()->value;
3238}
3239
3240} // namespace spirv
3241} // namespace mlir
3242MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::StorageClassAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::StorageClassAttr>::id = {}; } }
3243namespace mlir {
3244namespace spirv {
3245namespace detail {
3246struct VendorAttrStorage : public ::mlir::AttributeStorage {
3247 using KeyTy = std::tuple<::mlir::spirv::Vendor>;
3248 VendorAttrStorage(::mlir::spirv::Vendor value) : value(value) {}
3249
3250 KeyTy getAsKey() const {
3251 return KeyTy(value);
3252 }
3253
3254 bool operator==(const KeyTy &tblgenKey) const {
3255 return (value == std::get<0>(tblgenKey));
3256 }
3257
3258 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
3259 return ::llvm::hash_combine(std::get<0>(tblgenKey));
3260 }
3261
3262 static VendorAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
3263 auto value = std::get<0>(tblgenKey);
3264 return new (allocator.allocate<VendorAttrStorage>()) VendorAttrStorage(value);
3265 }
3266
3267 ::mlir::spirv::Vendor value;
3268};
3269} // namespace detail
3270VendorAttr VendorAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Vendor value) {
3271 return Base::get(context, value);
3272}
3273
3274::mlir::Attribute VendorAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
3275 ::mlir::Builder odsBuilder(odsParser.getContext());
3276 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
3277 (void) odsLoc;
3278 ::mlir::FailureOr<::mlir::spirv::Vendor> _result_value;
3279 // Parse literal '<'
3280 if (odsParser.parseLess()) return {};
3281
3282 // Parse variable 'value'
3283 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Vendor> {
3284 auto loc = odsParser.getCurrentLocation();
3285 ::llvm::StringRef enumKeyword;
3286 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
3287 return ::mlir::failure();
3288 auto maybeEnum = ::mlir::spirv::symbolizeVendor(enumKeyword);
3289 if (maybeEnum)
3290 return *maybeEnum;
3291 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::Vendor" << " to be one of: " << "AMD" << ", " << "Apple" << ", " << "ARM" << ", " << "Imagination" << ", " << "Intel" << ", " << "NVIDIA" << ", " << "Qualcomm" << ", " << "SwiftShader" << ", " << "Unknown")};
3292 }();
3293 if (::mlir::failed(_result_value)) {
3294 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_VendorAttr parameter 'value' which is to be a `::mlir::spirv::Vendor`");
3295 return {};
3296 }
3297 // Parse literal '>'
3298 if (odsParser.parseGreater()) return {};
3299 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 3299, __extension__ __PRETTY_FUNCTION__));
3300 return VendorAttr::get(odsParser.getContext(),
3301 ::mlir::spirv::Vendor((*_result_value)));
3302}
3303
3304void VendorAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3305 ::mlir::Builder odsBuilder(getContext());
3306 odsPrinter << "<";
3307 odsPrinter << stringifyVendor(getValue());
3308 odsPrinter << ">";
3309}
3310
3311::mlir::spirv::Vendor VendorAttr::getValue() const {
3312 return getImpl()->value;
3313}
3314
3315} // namespace spirv
3316} // namespace mlir
3317MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::VendorAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::VendorAttr>::id = {}; } }
3318namespace mlir {
3319namespace spirv {
3320namespace detail {
3321struct VersionAttrStorage : public ::mlir::AttributeStorage {
3322 using KeyTy = std::tuple<::mlir::spirv::Version>;
3323 VersionAttrStorage(::mlir::spirv::Version value) : value(value) {}
3324
3325 KeyTy getAsKey() const {
3326 return KeyTy(value);
3327 }
3328
3329 bool operator==(const KeyTy &tblgenKey) const {
3330 return (value == std::get<0>(tblgenKey));
3331 }
3332
3333 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
3334 return ::llvm::hash_combine(std::get<0>(tblgenKey));
3335 }
3336
3337 static VersionAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
3338 auto value = std::get<0>(tblgenKey);
3339 return new (allocator.allocate<VersionAttrStorage>()) VersionAttrStorage(value);
3340 }
3341
3342 ::mlir::spirv::Version value;
3343};
3344} // namespace detail
3345VersionAttr VersionAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Version value) {
3346 return Base::get(context, value);
3347}
3348
3349::mlir::Attribute VersionAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
3350 ::mlir::Builder odsBuilder(odsParser.getContext());
3351 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
3352 (void) odsLoc;
3353 ::mlir::FailureOr<::mlir::spirv::Version> _result_value;
3354 // Parse literal '<'
3355 if (odsParser.parseLess()) return {};
3356
3357 // Parse variable 'value'
3358 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Version> {
3359 auto loc = odsParser.getCurrentLocation();
3360 ::llvm::StringRef enumKeyword;
3361 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
3362 return ::mlir::failure();
3363 auto maybeEnum = ::mlir::spirv::symbolizeVersion(enumKeyword);
3364 if (maybeEnum)
3365 return *maybeEnum;
3366 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::Version" << " to be one of: " << "v1.0" << ", " << "v1.1" << ", " << "v1.2" << ", " << "v1.3" << ", " << "v1.4" << ", " << "v1.5" << ", " << "v1.6")};
3367 }();
3368 if (::mlir::failed(_result_value)) {
3369 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_VersionAttr parameter 'value' which is to be a `::mlir::spirv::Version`");
3370 return {};
3371 }
3372 // Parse literal '>'
3373 if (odsParser.parseGreater()) return {};
3374 assert(::mlir::succeeded(_result_value))(static_cast <bool> (::mlir::succeeded(_result_value)) ?
void (0) : __assert_fail ("::mlir::succeeded(_result_value)"
, "tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc"
, 3374, __extension__ __PRETTY_FUNCTION__));
3375 return VersionAttr::get(odsParser.getContext(),
3376 ::mlir::spirv::Version((*_result_value)));
3377}
3378
3379void VersionAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3380 ::mlir::Builder odsBuilder(getContext());
3381 odsPrinter << "<";
3382 odsPrinter << stringifyVersion(getValue());
3383 odsPrinter << ">";
3384}
3385
3386::mlir::spirv::Version VersionAttr::getValue() const {
3387 return getImpl()->value;
3388}
3389
3390} // namespace spirv
3391} // namespace mlir
3392MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::VersionAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::VersionAttr>::id = {}; } }
3393
3394#endif // GET_ATTRDEF_CLASSES
3395