Bug Summary

File:build/source/build-llvm/tools/clang/stage2-bins/tools/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVAttributes.cpp.inc
Warning:line 972, column 5
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-17/lib/clang/17 -D MLIR_CUDA_CONVERSIONS_ENABLED=1 -D MLIR_ROCM_CONVERSIONS_ENABLED=1 -D _DEBUG -D _GLIBCXX_ASSERTIONS -D _GNU_SOURCE -D _LIBCPP_ENABLE_ASSERTIONS -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-17/lib/clang/17/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 1683717183 -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-2023-05-10-133810-16478-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::ClientAPIAttr,
17::mlir::spirv::CooperativeMatrixPropertiesNVAttr,
18::mlir::spirv::DecorationAttr,
19::mlir::spirv::ImageDepthInfoAttr,
20::mlir::spirv::DeviceTypeAttr,
21::mlir::spirv::DimAttr,
22::mlir::spirv::EntryPointABIAttr,
23::mlir::spirv::ExecutionModeAttr,
24::mlir::spirv::ExecutionModelAttr,
25::mlir::spirv::ExtensionAttr,
26::mlir::spirv::FunctionControlAttr,
27::mlir::spirv::GroupOperationAttr,
28::mlir::spirv::ImageFormatAttr,
29::mlir::spirv::ImageOperandsAttr,
30::mlir::spirv::JointMatrixPropertiesINTELAttr,
31::mlir::spirv::LinkageTypeAttr,
32::mlir::spirv::LoopControlAttr,
33::mlir::spirv::MatrixLayoutAttr,
34::mlir::spirv::MemoryAccessAttr,
35::mlir::spirv::MemoryModelAttr,
36::mlir::spirv::MemorySemanticsAttr,
37::mlir::spirv::OpcodeAttr,
38::mlir::spirv::PackedVectorFormatAttr,
39::mlir::spirv::ResourceLimitsAttr,
40::mlir::spirv::ImageSamplerUseInfoAttr,
41::mlir::spirv::ImageSamplingInfoAttr,
42::mlir::spirv::ScopeAttr,
43::mlir::spirv::SelectionControlAttr,
44::mlir::spirv::StorageClassAttr,
45::mlir::spirv::VendorAttr,
46::mlir::spirv::VersionAttr
47
48#endif // GET_ATTRDEF_LIST
49
50#ifdef GET_ATTRDEF_CLASSES
51#undef GET_ATTRDEF_CLASSES
52
53static ::mlir::OptionalParseResult generatedAttributeParser(::mlir::AsmParser &parser, ::llvm::StringRef *mnemonic, ::mlir::Type type, ::mlir::Attribute &value) {
54 return ::mlir::AsmParser::KeywordSwitch<::mlir::OptionalParseResult>(parser)
55 .Case(::mlir::spirv::AddressingModelAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
56 value = ::mlir::spirv::AddressingModelAttr::parse(parser, type);
57 return ::mlir::success(!!value);
58 })
59 .Case(::mlir::spirv::ImageArrayedInfoAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
60 value = ::mlir::spirv::ImageArrayedInfoAttr::parse(parser, type);
61 return ::mlir::success(!!value);
62 })
63 .Case(::mlir::spirv::BuiltInAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
64 value = ::mlir::spirv::BuiltInAttr::parse(parser, type);
65 return ::mlir::success(!!value);
66 })
67 .Case(::mlir::spirv::CapabilityAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
68 value = ::mlir::spirv::CapabilityAttr::parse(parser, type);
69 return ::mlir::success(!!value);
70 })
71 .Case(::mlir::spirv::ClientAPIAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
72 value = ::mlir::spirv::ClientAPIAttr::parse(parser, type);
73 return ::mlir::success(!!value);
74 })
75 .Case(::mlir::spirv::CooperativeMatrixPropertiesNVAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
76 value = ::mlir::spirv::CooperativeMatrixPropertiesNVAttr::parse(parser, type);
77 return ::mlir::success(!!value);
78 })
79 .Case(::mlir::spirv::DecorationAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
80 value = ::mlir::spirv::DecorationAttr::parse(parser, type);
81 return ::mlir::success(!!value);
82 })
83 .Case(::mlir::spirv::ImageDepthInfoAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
84 value = ::mlir::spirv::ImageDepthInfoAttr::parse(parser, type);
85 return ::mlir::success(!!value);
86 })
87 .Case(::mlir::spirv::DeviceTypeAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
88 value = ::mlir::spirv::DeviceTypeAttr::parse(parser, type);
89 return ::mlir::success(!!value);
90 })
91 .Case(::mlir::spirv::DimAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
92 value = ::mlir::spirv::DimAttr::parse(parser, type);
93 return ::mlir::success(!!value);
94 })
95 .Case(::mlir::spirv::EntryPointABIAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
96 value = ::mlir::spirv::EntryPointABIAttr::parse(parser, type);
97 return ::mlir::success(!!value);
98 })
99 .Case(::mlir::spirv::ExecutionModeAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
100 value = ::mlir::spirv::ExecutionModeAttr::parse(parser, type);
101 return ::mlir::success(!!value);
102 })
103 .Case(::mlir::spirv::ExecutionModelAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
104 value = ::mlir::spirv::ExecutionModelAttr::parse(parser, type);
105 return ::mlir::success(!!value);
106 })
107 .Case(::mlir::spirv::ExtensionAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
108 value = ::mlir::spirv::ExtensionAttr::parse(parser, type);
109 return ::mlir::success(!!value);
110 })
111 .Case(::mlir::spirv::FunctionControlAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
112 value = ::mlir::spirv::FunctionControlAttr::parse(parser, type);
113 return ::mlir::success(!!value);
114 })
115 .Case(::mlir::spirv::GroupOperationAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
116 value = ::mlir::spirv::GroupOperationAttr::parse(parser, type);
117 return ::mlir::success(!!value);
118 })
119 .Case(::mlir::spirv::ImageFormatAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
120 value = ::mlir::spirv::ImageFormatAttr::parse(parser, type);
121 return ::mlir::success(!!value);
122 })
123 .Case(::mlir::spirv::ImageOperandsAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
124 value = ::mlir::spirv::ImageOperandsAttr::parse(parser, type);
125 return ::mlir::success(!!value);
126 })
127 .Case(::mlir::spirv::JointMatrixPropertiesINTELAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
128 value = ::mlir::spirv::JointMatrixPropertiesINTELAttr::parse(parser, type);
129 return ::mlir::success(!!value);
130 })
131 .Case(::mlir::spirv::LinkageTypeAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
132 value = ::mlir::spirv::LinkageTypeAttr::parse(parser, type);
133 return ::mlir::success(!!value);
134 })
135 .Case(::mlir::spirv::LoopControlAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
136 value = ::mlir::spirv::LoopControlAttr::parse(parser, type);
137 return ::mlir::success(!!value);
138 })
139 .Case(::mlir::spirv::MatrixLayoutAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
140 value = ::mlir::spirv::MatrixLayoutAttr::parse(parser, type);
141 return ::mlir::success(!!value);
142 })
143 .Case(::mlir::spirv::MemoryAccessAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
144 value = ::mlir::spirv::MemoryAccessAttr::parse(parser, type);
145 return ::mlir::success(!!value);
146 })
147 .Case(::mlir::spirv::MemoryModelAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
148 value = ::mlir::spirv::MemoryModelAttr::parse(parser, type);
149 return ::mlir::success(!!value);
150 })
151 .Case(::mlir::spirv::MemorySemanticsAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
152 value = ::mlir::spirv::MemorySemanticsAttr::parse(parser, type);
153 return ::mlir::success(!!value);
154 })
155 .Case(::mlir::spirv::OpcodeAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
156 value = ::mlir::spirv::OpcodeAttr::parse(parser, type);
157 return ::mlir::success(!!value);
158 })
159 .Case(::mlir::spirv::PackedVectorFormatAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
160 value = ::mlir::spirv::PackedVectorFormatAttr::parse(parser, type);
161 return ::mlir::success(!!value);
162 })
163 .Case(::mlir::spirv::ResourceLimitsAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
164 value = ::mlir::spirv::ResourceLimitsAttr::parse(parser, type);
165 return ::mlir::success(!!value);
166 })
167 .Case(::mlir::spirv::ImageSamplerUseInfoAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
168 value = ::mlir::spirv::ImageSamplerUseInfoAttr::parse(parser, type);
169 return ::mlir::success(!!value);
170 })
171 .Case(::mlir::spirv::ImageSamplingInfoAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
172 value = ::mlir::spirv::ImageSamplingInfoAttr::parse(parser, type);
173 return ::mlir::success(!!value);
174 })
175 .Case(::mlir::spirv::ScopeAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
176 value = ::mlir::spirv::ScopeAttr::parse(parser, type);
177 return ::mlir::success(!!value);
178 })
179 .Case(::mlir::spirv::SelectionControlAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
180 value = ::mlir::spirv::SelectionControlAttr::parse(parser, type);
181 return ::mlir::success(!!value);
182 })
183 .Case(::mlir::spirv::StorageClassAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
184 value = ::mlir::spirv::StorageClassAttr::parse(parser, type);
185 return ::mlir::success(!!value);
186 })
187 .Case(::mlir::spirv::VendorAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
188 value = ::mlir::spirv::VendorAttr::parse(parser, type);
189 return ::mlir::success(!!value);
190 })
191 .Case(::mlir::spirv::VersionAttr::getMnemonic(), [&](llvm::StringRef, llvm::SMLoc) {
192 value = ::mlir::spirv::VersionAttr::parse(parser, type);
193 return ::mlir::success(!!value);
194 })
195 .Default([&](llvm::StringRef keyword, llvm::SMLoc) {
196 *mnemonic = keyword;
197 return std::nullopt;
198 });
199}
200
201static ::mlir::LogicalResult generatedAttributePrinter(::mlir::Attribute def, ::mlir::AsmPrinter &printer) {
202 return ::llvm::TypeSwitch<::mlir::Attribute, ::mlir::LogicalResult>(def) .Case<::mlir::spirv::AddressingModelAttr>([&](auto t) {
203 printer << ::mlir::spirv::AddressingModelAttr::getMnemonic();
204t.print(printer);
205 return ::mlir::success();
206 })
207 .Case<::mlir::spirv::ImageArrayedInfoAttr>([&](auto t) {
208 printer << ::mlir::spirv::ImageArrayedInfoAttr::getMnemonic();
209t.print(printer);
210 return ::mlir::success();
211 })
212 .Case<::mlir::spirv::BuiltInAttr>([&](auto t) {
213 printer << ::mlir::spirv::BuiltInAttr::getMnemonic();
214t.print(printer);
215 return ::mlir::success();
216 })
217 .Case<::mlir::spirv::CapabilityAttr>([&](auto t) {
218 printer << ::mlir::spirv::CapabilityAttr::getMnemonic();
219t.print(printer);
220 return ::mlir::success();
221 })
222 .Case<::mlir::spirv::ClientAPIAttr>([&](auto t) {
223 printer << ::mlir::spirv::ClientAPIAttr::getMnemonic();
224t.print(printer);
225 return ::mlir::success();
226 })
227 .Case<::mlir::spirv::CooperativeMatrixPropertiesNVAttr>([&](auto t) {
228 printer << ::mlir::spirv::CooperativeMatrixPropertiesNVAttr::getMnemonic();
229t.print(printer);
230 return ::mlir::success();
231 })
232 .Case<::mlir::spirv::DecorationAttr>([&](auto t) {
233 printer << ::mlir::spirv::DecorationAttr::getMnemonic();
234t.print(printer);
235 return ::mlir::success();
236 })
237 .Case<::mlir::spirv::ImageDepthInfoAttr>([&](auto t) {
238 printer << ::mlir::spirv::ImageDepthInfoAttr::getMnemonic();
239t.print(printer);
240 return ::mlir::success();
241 })
242 .Case<::mlir::spirv::DeviceTypeAttr>([&](auto t) {
243 printer << ::mlir::spirv::DeviceTypeAttr::getMnemonic();
244t.print(printer);
245 return ::mlir::success();
246 })
247 .Case<::mlir::spirv::DimAttr>([&](auto t) {
248 printer << ::mlir::spirv::DimAttr::getMnemonic();
249t.print(printer);
250 return ::mlir::success();
251 })
252 .Case<::mlir::spirv::EntryPointABIAttr>([&](auto t) {
253 printer << ::mlir::spirv::EntryPointABIAttr::getMnemonic();
254t.print(printer);
255 return ::mlir::success();
256 })
257 .Case<::mlir::spirv::ExecutionModeAttr>([&](auto t) {
258 printer << ::mlir::spirv::ExecutionModeAttr::getMnemonic();
259t.print(printer);
260 return ::mlir::success();
261 })
262 .Case<::mlir::spirv::ExecutionModelAttr>([&](auto t) {
263 printer << ::mlir::spirv::ExecutionModelAttr::getMnemonic();
264t.print(printer);
265 return ::mlir::success();
266 })
267 .Case<::mlir::spirv::ExtensionAttr>([&](auto t) {
268 printer << ::mlir::spirv::ExtensionAttr::getMnemonic();
269t.print(printer);
270 return ::mlir::success();
271 })
272 .Case<::mlir::spirv::FunctionControlAttr>([&](auto t) {
273 printer << ::mlir::spirv::FunctionControlAttr::getMnemonic();
274t.print(printer);
275 return ::mlir::success();
276 })
277 .Case<::mlir::spirv::GroupOperationAttr>([&](auto t) {
278 printer << ::mlir::spirv::GroupOperationAttr::getMnemonic();
279t.print(printer);
280 return ::mlir::success();
281 })
282 .Case<::mlir::spirv::ImageFormatAttr>([&](auto t) {
283 printer << ::mlir::spirv::ImageFormatAttr::getMnemonic();
284t.print(printer);
285 return ::mlir::success();
286 })
287 .Case<::mlir::spirv::ImageOperandsAttr>([&](auto t) {
288 printer << ::mlir::spirv::ImageOperandsAttr::getMnemonic();
289t.print(printer);
290 return ::mlir::success();
291 })
292 .Case<::mlir::spirv::JointMatrixPropertiesINTELAttr>([&](auto t) {
293 printer << ::mlir::spirv::JointMatrixPropertiesINTELAttr::getMnemonic();
294t.print(printer);
295 return ::mlir::success();
296 })
297 .Case<::mlir::spirv::LinkageTypeAttr>([&](auto t) {
298 printer << ::mlir::spirv::LinkageTypeAttr::getMnemonic();
299t.print(printer);
300 return ::mlir::success();
301 })
302 .Case<::mlir::spirv::LoopControlAttr>([&](auto t) {
303 printer << ::mlir::spirv::LoopControlAttr::getMnemonic();
304t.print(printer);
305 return ::mlir::success();
306 })
307 .Case<::mlir::spirv::MatrixLayoutAttr>([&](auto t) {
308 printer << ::mlir::spirv::MatrixLayoutAttr::getMnemonic();
309t.print(printer);
310 return ::mlir::success();
311 })
312 .Case<::mlir::spirv::MemoryAccessAttr>([&](auto t) {
313 printer << ::mlir::spirv::MemoryAccessAttr::getMnemonic();
314t.print(printer);
315 return ::mlir::success();
316 })
317 .Case<::mlir::spirv::MemoryModelAttr>([&](auto t) {
318 printer << ::mlir::spirv::MemoryModelAttr::getMnemonic();
319t.print(printer);
320 return ::mlir::success();
321 })
322 .Case<::mlir::spirv::MemorySemanticsAttr>([&](auto t) {
323 printer << ::mlir::spirv::MemorySemanticsAttr::getMnemonic();
324t.print(printer);
325 return ::mlir::success();
326 })
327 .Case<::mlir::spirv::OpcodeAttr>([&](auto t) {
328 printer << ::mlir::spirv::OpcodeAttr::getMnemonic();
329t.print(printer);
330 return ::mlir::success();
331 })
332 .Case<::mlir::spirv::PackedVectorFormatAttr>([&](auto t) {
333 printer << ::mlir::spirv::PackedVectorFormatAttr::getMnemonic();
334t.print(printer);
335 return ::mlir::success();
336 })
337 .Case<::mlir::spirv::ResourceLimitsAttr>([&](auto t) {
338 printer << ::mlir::spirv::ResourceLimitsAttr::getMnemonic();
339t.print(printer);
340 return ::mlir::success();
341 })
342 .Case<::mlir::spirv::ImageSamplerUseInfoAttr>([&](auto t) {
343 printer << ::mlir::spirv::ImageSamplerUseInfoAttr::getMnemonic();
344t.print(printer);
345 return ::mlir::success();
346 })
347 .Case<::mlir::spirv::ImageSamplingInfoAttr>([&](auto t) {
348 printer << ::mlir::spirv::ImageSamplingInfoAttr::getMnemonic();
349t.print(printer);
350 return ::mlir::success();
351 })
352 .Case<::mlir::spirv::ScopeAttr>([&](auto t) {
353 printer << ::mlir::spirv::ScopeAttr::getMnemonic();
354t.print(printer);
355 return ::mlir::success();
356 })
357 .Case<::mlir::spirv::SelectionControlAttr>([&](auto t) {
358 printer << ::mlir::spirv::SelectionControlAttr::getMnemonic();
359t.print(printer);
360 return ::mlir::success();
361 })
362 .Case<::mlir::spirv::StorageClassAttr>([&](auto t) {
363 printer << ::mlir::spirv::StorageClassAttr::getMnemonic();
364t.print(printer);
365 return ::mlir::success();
366 })
367 .Case<::mlir::spirv::VendorAttr>([&](auto t) {
368 printer << ::mlir::spirv::VendorAttr::getMnemonic();
369t.print(printer);
370 return ::mlir::success();
371 })
372 .Case<::mlir::spirv::VersionAttr>([&](auto t) {
373 printer << ::mlir::spirv::VersionAttr::getMnemonic();
374t.print(printer);
375 return ::mlir::success();
376 })
377 .Default([](auto) { return ::mlir::failure(); });
378}
379
380namespace mlir {
381namespace spirv {
382namespace detail {
383struct AddressingModelAttrStorage : public ::mlir::AttributeStorage {
384 using KeyTy = std::tuple<::mlir::spirv::AddressingModel>;
385 AddressingModelAttrStorage(::mlir::spirv::AddressingModel value) : value(value) {}
386
387 KeyTy getAsKey() const {
388 return KeyTy(value);
389 }
390
391 bool operator==(const KeyTy &tblgenKey) const {
392 return (value == std::get<0>(tblgenKey));
393 }
394
395 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
396 return ::llvm::hash_combine(std::get<0>(tblgenKey));
397 }
398
399 static AddressingModelAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
400 auto value = std::get<0>(tblgenKey);
401 return new (allocator.allocate<AddressingModelAttrStorage>()) AddressingModelAttrStorage(value);
402 }
403
404 ::mlir::spirv::AddressingModel value;
405};
406} // namespace detail
407AddressingModelAttr AddressingModelAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::AddressingModel value) {
408 return Base::get(context, value);
409}
410
411::mlir::Attribute AddressingModelAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
412 ::mlir::Builder odsBuilder(odsParser.getContext());
413 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
414 (void) odsLoc;
415 ::mlir::FailureOr<::mlir::spirv::AddressingModel> _result_value;
416 // Parse literal '<'
417 if (odsParser.parseLess()) return {};
418
419 // Parse variable 'value'
420 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::AddressingModel> {
421 auto loc = odsParser.getCurrentLocation();
422 ::llvm::StringRef enumKeyword;
423 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
424 return ::mlir::failure();
425 auto maybeEnum = ::mlir::spirv::symbolizeAddressingModel(enumKeyword);
426 if (maybeEnum)
427 return *maybeEnum;
428 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::AddressingModel" << " to be one of: " << "Logical" << ", " << "Physical32" << ", " << "Physical64" << ", " << "PhysicalStorageBuffer64")};
429 }();
430 if (::mlir::failed(_result_value)) {
431 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_AddressingModelAttr parameter 'value' which is to be a `::mlir::spirv::AddressingModel`");
432 return {};
433 }
434 // Parse literal '>'
435 if (odsParser.parseGreater()) return {};
436 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"
, 436, __extension__ __PRETTY_FUNCTION__));
437 return AddressingModelAttr::get(odsParser.getContext(),
438 ::mlir::spirv::AddressingModel((*_result_value)));
439}
440
441void AddressingModelAttr::print(::mlir::AsmPrinter &odsPrinter) const {
442 ::mlir::Builder odsBuilder(getContext());
443 odsPrinter << "<";
444 odsPrinter << stringifyAddressingModel(getValue());
445 odsPrinter << ">";
446}
447
448::mlir::spirv::AddressingModel AddressingModelAttr::getValue() const {
449 return getImpl()->value;
450}
451
452} // namespace spirv
453} // namespace mlir
454MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::AddressingModelAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::AddressingModelAttr>::id = {}; } }
455namespace mlir {
456namespace spirv {
457namespace detail {
458struct ImageArrayedInfoAttrStorage : public ::mlir::AttributeStorage {
459 using KeyTy = std::tuple<::mlir::spirv::ImageArrayedInfo>;
460 ImageArrayedInfoAttrStorage(::mlir::spirv::ImageArrayedInfo value) : value(value) {}
461
462 KeyTy getAsKey() const {
463 return KeyTy(value);
464 }
465
466 bool operator==(const KeyTy &tblgenKey) const {
467 return (value == std::get<0>(tblgenKey));
468 }
469
470 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
471 return ::llvm::hash_combine(std::get<0>(tblgenKey));
472 }
473
474 static ImageArrayedInfoAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
475 auto value = std::get<0>(tblgenKey);
476 return new (allocator.allocate<ImageArrayedInfoAttrStorage>()) ImageArrayedInfoAttrStorage(value);
477 }
478
479 ::mlir::spirv::ImageArrayedInfo value;
480};
481} // namespace detail
482ImageArrayedInfoAttr ImageArrayedInfoAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ImageArrayedInfo value) {
483 return Base::get(context, value);
484}
485
486::mlir::Attribute ImageArrayedInfoAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
487 ::mlir::Builder odsBuilder(odsParser.getContext());
488 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
489 (void) odsLoc;
490 ::mlir::FailureOr<::mlir::spirv::ImageArrayedInfo> _result_value;
491 // Parse literal '<'
492 if (odsParser.parseLess()) return {};
493
494 // Parse variable 'value'
495 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ImageArrayedInfo> {
496 auto loc = odsParser.getCurrentLocation();
497 ::llvm::StringRef enumKeyword;
498 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
499 return ::mlir::failure();
500 auto maybeEnum = ::mlir::spirv::symbolizeImageArrayedInfo(enumKeyword);
501 if (maybeEnum)
502 return *maybeEnum;
503 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ImageArrayedInfo" << " to be one of: " << "NonArrayed" << ", " << "Arrayed")};
504 }();
505 if (::mlir::failed(_result_value)) {
506 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ArrayedAttr parameter 'value' which is to be a `::mlir::spirv::ImageArrayedInfo`");
507 return {};
508 }
509 // Parse literal '>'
510 if (odsParser.parseGreater()) return {};
511 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"
, 511, __extension__ __PRETTY_FUNCTION__));
512 return ImageArrayedInfoAttr::get(odsParser.getContext(),
513 ::mlir::spirv::ImageArrayedInfo((*_result_value)));
514}
515
516void ImageArrayedInfoAttr::print(::mlir::AsmPrinter &odsPrinter) const {
517 ::mlir::Builder odsBuilder(getContext());
518 odsPrinter << "<";
519 odsPrinter << stringifyImageArrayedInfo(getValue());
520 odsPrinter << ">";
521}
522
523::mlir::spirv::ImageArrayedInfo ImageArrayedInfoAttr::getValue() const {
524 return getImpl()->value;
525}
526
527} // namespace spirv
528} // namespace mlir
529MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ImageArrayedInfoAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ImageArrayedInfoAttr>::id = {}; } }
530namespace mlir {
531namespace spirv {
532namespace detail {
533struct BuiltInAttrStorage : public ::mlir::AttributeStorage {
534 using KeyTy = std::tuple<::mlir::spirv::BuiltIn>;
535 BuiltInAttrStorage(::mlir::spirv::BuiltIn value) : value(value) {}
536
537 KeyTy getAsKey() const {
538 return KeyTy(value);
539 }
540
541 bool operator==(const KeyTy &tblgenKey) const {
542 return (value == std::get<0>(tblgenKey));
543 }
544
545 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
546 return ::llvm::hash_combine(std::get<0>(tblgenKey));
547 }
548
549 static BuiltInAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
550 auto value = std::get<0>(tblgenKey);
551 return new (allocator.allocate<BuiltInAttrStorage>()) BuiltInAttrStorage(value);
552 }
553
554 ::mlir::spirv::BuiltIn value;
555};
556} // namespace detail
557BuiltInAttr BuiltInAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::BuiltIn value) {
558 return Base::get(context, value);
559}
560
561::mlir::Attribute BuiltInAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
562 ::mlir::Builder odsBuilder(odsParser.getContext());
563 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
564 (void) odsLoc;
565 ::mlir::FailureOr<::mlir::spirv::BuiltIn> _result_value;
566 // Parse literal '<'
567 if (odsParser.parseLess()) return {};
568
569 // Parse variable 'value'
570 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::BuiltIn> {
571 auto loc = odsParser.getCurrentLocation();
572 ::llvm::StringRef enumKeyword;
573 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
574 return ::mlir::failure();
575 auto maybeEnum = ::mlir::spirv::symbolizeBuiltIn(enumKeyword);
576 if (maybeEnum)
577 return *maybeEnum;
578 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")};
579 }();
580 if (::mlir::failed(_result_value)) {
581 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_BuiltInAttr parameter 'value' which is to be a `::mlir::spirv::BuiltIn`");
582 return {};
583 }
584 // Parse literal '>'
585 if (odsParser.parseGreater()) return {};
586 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"
, 586, __extension__ __PRETTY_FUNCTION__));
587 return BuiltInAttr::get(odsParser.getContext(),
588 ::mlir::spirv::BuiltIn((*_result_value)));
589}
590
591void BuiltInAttr::print(::mlir::AsmPrinter &odsPrinter) const {
592 ::mlir::Builder odsBuilder(getContext());
593 odsPrinter << "<";
594 odsPrinter << stringifyBuiltIn(getValue());
595 odsPrinter << ">";
596}
597
598::mlir::spirv::BuiltIn BuiltInAttr::getValue() const {
599 return getImpl()->value;
600}
601
602} // namespace spirv
603} // namespace mlir
604MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::BuiltInAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::BuiltInAttr>::id = {}; } }
605namespace mlir {
606namespace spirv {
607namespace detail {
608struct CapabilityAttrStorage : public ::mlir::AttributeStorage {
609 using KeyTy = std::tuple<::mlir::spirv::Capability>;
610 CapabilityAttrStorage(::mlir::spirv::Capability value) : value(value) {}
611
612 KeyTy getAsKey() const {
613 return KeyTy(value);
614 }
615
616 bool operator==(const KeyTy &tblgenKey) const {
617 return (value == std::get<0>(tblgenKey));
618 }
619
620 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
621 return ::llvm::hash_combine(std::get<0>(tblgenKey));
622 }
623
624 static CapabilityAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
625 auto value = std::get<0>(tblgenKey);
626 return new (allocator.allocate<CapabilityAttrStorage>()) CapabilityAttrStorage(value);
627 }
628
629 ::mlir::spirv::Capability value;
630};
631} // namespace detail
632CapabilityAttr CapabilityAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Capability value) {
633 return Base::get(context, value);
634}
635
636::mlir::Attribute CapabilityAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
637 ::mlir::Builder odsBuilder(odsParser.getContext());
638 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
639 (void) odsLoc;
640 ::mlir::FailureOr<::mlir::spirv::Capability> _result_value;
641 // Parse literal '<'
642 if (odsParser.parseLess()) return {};
643
644 // Parse variable 'value'
645 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Capability> {
646 auto loc = odsParser.getCurrentLocation();
647 ::llvm::StringRef enumKeyword;
648 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
649 return ::mlir::failure();
650 auto maybeEnum = ::mlir::spirv::symbolizeCapability(enumKeyword);
651 if (maybeEnum)
652 return *maybeEnum;
653 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" << ", " << "Bfloat16ConversionINTEL")};
654 }();
655 if (::mlir::failed(_result_value)) {
656 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CapabilityAttr parameter 'value' which is to be a `::mlir::spirv::Capability`");
657 return {};
658 }
659 // Parse literal '>'
660 if (odsParser.parseGreater()) return {};
661 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"
, 661, __extension__ __PRETTY_FUNCTION__));
662 return CapabilityAttr::get(odsParser.getContext(),
663 ::mlir::spirv::Capability((*_result_value)));
664}
665
666void CapabilityAttr::print(::mlir::AsmPrinter &odsPrinter) const {
667 ::mlir::Builder odsBuilder(getContext());
668 odsPrinter << "<";
669 odsPrinter << stringifyCapability(getValue());
670 odsPrinter << ">";
671}
672
673::mlir::spirv::Capability CapabilityAttr::getValue() const {
674 return getImpl()->value;
675}
676
677} // namespace spirv
678} // namespace mlir
679MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::CapabilityAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::CapabilityAttr>::id = {}; } }
680namespace mlir {
681namespace spirv {
682namespace detail {
683struct ClientAPIAttrStorage : public ::mlir::AttributeStorage {
684 using KeyTy = std::tuple<::mlir::spirv::ClientAPI>;
685 ClientAPIAttrStorage(::mlir::spirv::ClientAPI value) : value(value) {}
686
687 KeyTy getAsKey() const {
688 return KeyTy(value);
689 }
690
691 bool operator==(const KeyTy &tblgenKey) const {
692 return (value == std::get<0>(tblgenKey));
693 }
694
695 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
696 return ::llvm::hash_combine(std::get<0>(tblgenKey));
697 }
698
699 static ClientAPIAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
700 auto value = std::get<0>(tblgenKey);
701 return new (allocator.allocate<ClientAPIAttrStorage>()) ClientAPIAttrStorage(value);
702 }
703
704 ::mlir::spirv::ClientAPI value;
705};
706} // namespace detail
707ClientAPIAttr ClientAPIAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ClientAPI value) {
708 return Base::get(context, value);
709}
710
711::mlir::Attribute ClientAPIAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
712 ::mlir::Builder odsBuilder(odsParser.getContext());
713 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
714 (void) odsLoc;
715 ::mlir::FailureOr<::mlir::spirv::ClientAPI> _result_value;
716 // Parse literal '<'
717 if (odsParser.parseLess()) return {};
718
719 // Parse variable 'value'
720 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ClientAPI> {
721 auto loc = odsParser.getCurrentLocation();
722 ::llvm::StringRef enumKeyword;
723 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
724 return ::mlir::failure();
725 auto maybeEnum = ::mlir::spirv::symbolizeClientAPI(enumKeyword);
726 if (maybeEnum)
727 return *maybeEnum;
728 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ClientAPI" << " to be one of: " << "Metal" << ", " << "OpenCL" << ", " << "Vulkan" << ", " << "WebGPU" << ", " << "Unknown")};
729 }();
730 if (::mlir::failed(_result_value)) {
731 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ClientAPIAttr parameter 'value' which is to be a `::mlir::spirv::ClientAPI`");
732 return {};
733 }
734 // Parse literal '>'
735 if (odsParser.parseGreater()) return {};
736 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"
, 736, __extension__ __PRETTY_FUNCTION__));
737 return ClientAPIAttr::get(odsParser.getContext(),
738 ::mlir::spirv::ClientAPI((*_result_value)));
739}
740
741void ClientAPIAttr::print(::mlir::AsmPrinter &odsPrinter) const {
742 ::mlir::Builder odsBuilder(getContext());
743 odsPrinter << "<";
744 odsPrinter << stringifyClientAPI(getValue());
745 odsPrinter << ">";
746}
747
748::mlir::spirv::ClientAPI ClientAPIAttr::getValue() const {
749 return getImpl()->value;
750}
751
752} // namespace spirv
753} // namespace mlir
754MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ClientAPIAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ClientAPIAttr>::id = {}; } }
755namespace mlir {
756namespace spirv {
757namespace detail {
758struct CooperativeMatrixPropertiesNVAttrStorage : public ::mlir::AttributeStorage {
759 using KeyTy = std::tuple<int, int, int, mlir::Type, mlir::Type, mlir::Type, mlir::Type, mlir::spirv::ScopeAttr>;
760 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) {}
761
762 KeyTy getAsKey() const {
763 return KeyTy(m_size, n_size, k_size, a_type, b_type, c_type, result_type, scope);
764 }
765
766 bool operator==(const KeyTy &tblgenKey) const {
767 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));
768 }
769
770 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
771 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));
772 }
773
774 static CooperativeMatrixPropertiesNVAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
775 auto m_size = std::get<0>(tblgenKey);
776 auto n_size = std::get<1>(tblgenKey);
777 auto k_size = std::get<2>(tblgenKey);
778 auto a_type = std::get<3>(tblgenKey);
779 auto b_type = std::get<4>(tblgenKey);
780 auto c_type = std::get<5>(tblgenKey);
781 auto result_type = std::get<6>(tblgenKey);
782 auto scope = std::get<7>(tblgenKey);
783 return new (allocator.allocate<CooperativeMatrixPropertiesNVAttrStorage>()) CooperativeMatrixPropertiesNVAttrStorage(m_size, n_size, k_size, a_type, b_type, c_type, result_type, scope);
784 }
785
786 int m_size;
787 int n_size;
788 int k_size;
789 mlir::Type a_type;
790 mlir::Type b_type;
791 mlir::Type c_type;
792 mlir::Type result_type;
793 mlir::spirv::ScopeAttr scope;
794};
795} // namespace detail
796CooperativeMatrixPropertiesNVAttr 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) {
797 return Base::get(context, m_size, n_size, k_size, a_type, b_type, c_type, result_type, scope);
798}
799
800::mlir::Attribute CooperativeMatrixPropertiesNVAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
801 ::mlir::Builder odsBuilder(odsParser.getContext());
802 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
803 (void) odsLoc;
804 ::mlir::FailureOr<int> _result_m_size;
805 ::mlir::FailureOr<int> _result_n_size;
806 ::mlir::FailureOr<int> _result_k_size;
807 ::mlir::FailureOr<mlir::Type> _result_a_type;
808 ::mlir::FailureOr<mlir::Type> _result_b_type;
809 ::mlir::FailureOr<mlir::Type> _result_c_type;
810 ::mlir::FailureOr<mlir::Type> _result_result_type;
811 ::mlir::FailureOr<mlir::spirv::ScopeAttr> _result_scope;
812 // Parse literal '<'
813 if (odsParser.parseLess()) return {};
814 // Parse parameter struct
815 bool _seen_m_size = false;
816 bool _seen_n_size = false;
817 bool _seen_k_size = false;
818 bool _seen_a_type = false;
819 bool _seen_b_type = false;
820 bool _seen_c_type = false;
821 bool _seen_result_type = false;
822 bool _seen_scope = false;
823 {
824 const auto _loop_body = [&](::llvm::StringRef _paramKey) -> bool {
825 // Parse literal '='
826 if (odsParser.parseEqual()) return {};
827 if (!_seen_m_size && _paramKey == "m_size") {
828 _seen_m_size = true;
829
830 // Parse variable 'm_size'
831 _result_m_size = ::mlir::FieldParser<int>::parse(odsParser);
832 if (::mlir::failed(_result_m_size)) {
833 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'm_size' which is to be a `int`");
834 return {};
835 }
836 } else if (!_seen_n_size && _paramKey == "n_size") {
837 _seen_n_size = true;
838
839 // Parse variable 'n_size'
840 _result_n_size = ::mlir::FieldParser<int>::parse(odsParser);
841 if (::mlir::failed(_result_n_size)) {
842 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'n_size' which is to be a `int`");
843 return {};
844 }
845 } else if (!_seen_k_size && _paramKey == "k_size") {
846 _seen_k_size = true;
847
848 // Parse variable 'k_size'
849 _result_k_size = ::mlir::FieldParser<int>::parse(odsParser);
850 if (::mlir::failed(_result_k_size)) {
851 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'k_size' which is to be a `int`");
852 return {};
853 }
854 } else if (!_seen_a_type && _paramKey == "a_type") {
855 _seen_a_type = true;
856
857 // Parse variable 'a_type'
858 _result_a_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
859 if (::mlir::failed(_result_a_type)) {
860 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'a_type' which is to be a `mlir::Type`");
861 return {};
862 }
863 } else if (!_seen_b_type && _paramKey == "b_type") {
864 _seen_b_type = true;
865
866 // Parse variable 'b_type'
867 _result_b_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
868 if (::mlir::failed(_result_b_type)) {
869 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'b_type' which is to be a `mlir::Type`");
870 return {};
871 }
872 } else if (!_seen_c_type && _paramKey == "c_type") {
873 _seen_c_type = true;
874
875 // Parse variable 'c_type'
876 _result_c_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
877 if (::mlir::failed(_result_c_type)) {
878 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'c_type' which is to be a `mlir::Type`");
879 return {};
880 }
881 } else if (!_seen_result_type && _paramKey == "result_type") {
882 _seen_result_type = true;
883
884 // Parse variable 'result_type'
885 _result_result_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
886 if (::mlir::failed(_result_result_type)) {
887 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'result_type' which is to be a `mlir::Type`");
888 return {};
889 }
890 } else if (!_seen_scope && _paramKey == "scope") {
891 _seen_scope = true;
892
893 // Parse variable 'scope'
894 _result_scope = ::mlir::FieldParser<mlir::spirv::ScopeAttr>::parse(odsParser);
895 if (::mlir::failed(_result_scope)) {
896 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_CooperativeMatrixPropertiesNVAttr parameter 'scope' which is to be a `mlir::spirv::ScopeAttr`");
897 return {};
898 }
899 } else {
900 odsParser.emitError(odsParser.getCurrentLocation(), "duplicate or unknown struct parameter name: ") << _paramKey;
901 return {};
902 }
903 return true;
904 };
905 for (unsigned odsStructIndex = 0; odsStructIndex < 8; ++odsStructIndex) {
906 ::llvm::StringRef _paramKey;
907 if (odsParser.parseKeyword(&_paramKey)) {
908 odsParser.emitError(odsParser.getCurrentLocation(),
909 "expected a parameter name in struct");
910 return {};
911 }
912 if (!_loop_body(_paramKey)) return {};
913 if ((odsStructIndex != 8 - 1) && odsParser.parseComma())
914 return {};
915 }
916 }
917 // Parse literal '>'
918 if (odsParser.parseGreater()) return {};
919 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"
, 919, __extension__ __PRETTY_FUNCTION__));
920 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"
, 920, __extension__ __PRETTY_FUNCTION__));
921 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"
, 921, __extension__ __PRETTY_FUNCTION__));
922 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"
, 922, __extension__ __PRETTY_FUNCTION__));
923 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"
, 923, __extension__ __PRETTY_FUNCTION__));
924 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"
, 924, __extension__ __PRETTY_FUNCTION__));
925 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"
, 925, __extension__ __PRETTY_FUNCTION__));
926 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"
, 926, __extension__ __PRETTY_FUNCTION__));
927 return CooperativeMatrixPropertiesNVAttr::get(odsParser.getContext(),
928 int((*_result_m_size)),
929 int((*_result_n_size)),
930 int((*_result_k_size)),
931 mlir::Type((*_result_a_type)),
932 mlir::Type((*_result_b_type)),
933 mlir::Type((*_result_c_type)),
934 mlir::Type((*_result_result_type)),
935 mlir::spirv::ScopeAttr((*_result_scope)));
936}
937
938void CooperativeMatrixPropertiesNVAttr::print(::mlir::AsmPrinter &odsPrinter) const {
939 ::mlir::Builder odsBuilder(getContext());
940 odsPrinter << "<";
941 {
942 bool _firstPrinted = true;
943 if (!_firstPrinted) odsPrinter << ", ";
944 _firstPrinted = false;
945 odsPrinter << "m_size = ";
946 odsPrinter.printStrippedAttrOrType(getMSize());
947 if (!_firstPrinted) odsPrinter << ", ";
948 _firstPrinted = false;
949 odsPrinter << "n_size = ";
950 odsPrinter.printStrippedAttrOrType(getNSize());
951 if (!_firstPrinted) odsPrinter << ", ";
952 _firstPrinted = false;
953 odsPrinter << "k_size = ";
954 odsPrinter.printStrippedAttrOrType(getKSize());
955 if (!_firstPrinted) odsPrinter << ", ";
956 _firstPrinted = false;
957 odsPrinter << "a_type = ";
958 odsPrinter.printStrippedAttrOrType(getAType());
959 if (!_firstPrinted) odsPrinter << ", ";
960 _firstPrinted = false;
961 odsPrinter << "b_type = ";
962 odsPrinter.printStrippedAttrOrType(getBType());
963 if (!_firstPrinted) odsPrinter << ", ";
964 _firstPrinted = false;
965 odsPrinter << "c_type = ";
966 odsPrinter.printStrippedAttrOrType(getCType());
967 if (!_firstPrinted) odsPrinter << ", ";
968 _firstPrinted = false;
969 odsPrinter << "result_type = ";
970 odsPrinter.printStrippedAttrOrType(getResultType());
971 if (!_firstPrinted) odsPrinter << ", ";
972 _firstPrinted = false;
Value stored to '_firstPrinted' is never read
973 odsPrinter << "scope = ";
974 odsPrinter.printStrippedAttrOrType(getScope());
975 }
976 odsPrinter << ">";
977}
978
979int CooperativeMatrixPropertiesNVAttr::getMSize() const {
980 return getImpl()->m_size;
981}
982
983int CooperativeMatrixPropertiesNVAttr::getNSize() const {
984 return getImpl()->n_size;
985}
986
987int CooperativeMatrixPropertiesNVAttr::getKSize() const {
988 return getImpl()->k_size;
989}
990
991mlir::Type CooperativeMatrixPropertiesNVAttr::getAType() const {
992 return getImpl()->a_type;
993}
994
995mlir::Type CooperativeMatrixPropertiesNVAttr::getBType() const {
996 return getImpl()->b_type;
997}
998
999mlir::Type CooperativeMatrixPropertiesNVAttr::getCType() const {
1000 return getImpl()->c_type;
1001}
1002
1003mlir::Type CooperativeMatrixPropertiesNVAttr::getResultType() const {
1004 return getImpl()->result_type;
1005}
1006
1007mlir::spirv::ScopeAttr CooperativeMatrixPropertiesNVAttr::getScope() const {
1008 return getImpl()->scope;
1009}
1010
1011} // namespace spirv
1012} // namespace mlir
1013MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::CooperativeMatrixPropertiesNVAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::CooperativeMatrixPropertiesNVAttr>::id
= {}; } }
1014namespace mlir {
1015namespace spirv {
1016namespace detail {
1017struct DecorationAttrStorage : public ::mlir::AttributeStorage {
1018 using KeyTy = std::tuple<::mlir::spirv::Decoration>;
1019 DecorationAttrStorage(::mlir::spirv::Decoration value) : value(value) {}
1020
1021 KeyTy getAsKey() const {
1022 return KeyTy(value);
1023 }
1024
1025 bool operator==(const KeyTy &tblgenKey) const {
1026 return (value == std::get<0>(tblgenKey));
1027 }
1028
1029 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1030 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1031 }
1032
1033 static DecorationAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1034 auto value = std::get<0>(tblgenKey);
1035 return new (allocator.allocate<DecorationAttrStorage>()) DecorationAttrStorage(value);
1036 }
1037
1038 ::mlir::spirv::Decoration value;
1039};
1040} // namespace detail
1041DecorationAttr DecorationAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Decoration value) {
1042 return Base::get(context, value);
1043}
1044
1045::mlir::Attribute DecorationAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1046 ::mlir::Builder odsBuilder(odsParser.getContext());
1047 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1048 (void) odsLoc;
1049 ::mlir::FailureOr<::mlir::spirv::Decoration> _result_value;
1050 // Parse literal '<'
1051 if (odsParser.parseLess()) return {};
1052
1053 // Parse variable 'value'
1054 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Decoration> {
1055 auto loc = odsParser.getCurrentLocation();
1056 ::llvm::StringRef enumKeyword;
1057 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1058 return ::mlir::failure();
1059 auto maybeEnum = ::mlir::spirv::symbolizeDecoration(enumKeyword);
1060 if (maybeEnum)
1061 return *maybeEnum;
1062 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")};
1063 }();
1064 if (::mlir::failed(_result_value)) {
1065 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_DecorationAttr parameter 'value' which is to be a `::mlir::spirv::Decoration`");
1066 return {};
1067 }
1068 // Parse literal '>'
1069 if (odsParser.parseGreater()) return {};
1070 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"
, 1070, __extension__ __PRETTY_FUNCTION__));
1071 return DecorationAttr::get(odsParser.getContext(),
1072 ::mlir::spirv::Decoration((*_result_value)));
1073}
1074
1075void DecorationAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1076 ::mlir::Builder odsBuilder(getContext());
1077 odsPrinter << "<";
1078 odsPrinter << stringifyDecoration(getValue());
1079 odsPrinter << ">";
1080}
1081
1082::mlir::spirv::Decoration DecorationAttr::getValue() const {
1083 return getImpl()->value;
1084}
1085
1086} // namespace spirv
1087} // namespace mlir
1088MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::DecorationAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::DecorationAttr>::id = {}; } }
1089namespace mlir {
1090namespace spirv {
1091namespace detail {
1092struct ImageDepthInfoAttrStorage : public ::mlir::AttributeStorage {
1093 using KeyTy = std::tuple<::mlir::spirv::ImageDepthInfo>;
1094 ImageDepthInfoAttrStorage(::mlir::spirv::ImageDepthInfo value) : value(value) {}
1095
1096 KeyTy getAsKey() const {
1097 return KeyTy(value);
1098 }
1099
1100 bool operator==(const KeyTy &tblgenKey) const {
1101 return (value == std::get<0>(tblgenKey));
1102 }
1103
1104 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1105 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1106 }
1107
1108 static ImageDepthInfoAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1109 auto value = std::get<0>(tblgenKey);
1110 return new (allocator.allocate<ImageDepthInfoAttrStorage>()) ImageDepthInfoAttrStorage(value);
1111 }
1112
1113 ::mlir::spirv::ImageDepthInfo value;
1114};
1115} // namespace detail
1116ImageDepthInfoAttr ImageDepthInfoAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ImageDepthInfo value) {
1117 return Base::get(context, value);
1118}
1119
1120::mlir::Attribute ImageDepthInfoAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1121 ::mlir::Builder odsBuilder(odsParser.getContext());
1122 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1123 (void) odsLoc;
1124 ::mlir::FailureOr<::mlir::spirv::ImageDepthInfo> _result_value;
1125 // Parse literal '<'
1126 if (odsParser.parseLess()) return {};
1127
1128 // Parse variable 'value'
1129 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ImageDepthInfo> {
1130 auto loc = odsParser.getCurrentLocation();
1131 ::llvm::StringRef enumKeyword;
1132 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1133 return ::mlir::failure();
1134 auto maybeEnum = ::mlir::spirv::symbolizeImageDepthInfo(enumKeyword);
1135 if (maybeEnum)
1136 return *maybeEnum;
1137 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ImageDepthInfo" << " to be one of: " << "NoDepth" << ", " << "IsDepth" << ", " << "DepthUnknown")};
1138 }();
1139 if (::mlir::failed(_result_value)) {
1140 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_DepthAttr parameter 'value' which is to be a `::mlir::spirv::ImageDepthInfo`");
1141 return {};
1142 }
1143 // Parse literal '>'
1144 if (odsParser.parseGreater()) return {};
1145 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"
, 1145, __extension__ __PRETTY_FUNCTION__));
1146 return ImageDepthInfoAttr::get(odsParser.getContext(),
1147 ::mlir::spirv::ImageDepthInfo((*_result_value)));
1148}
1149
1150void ImageDepthInfoAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1151 ::mlir::Builder odsBuilder(getContext());
1152 odsPrinter << "<";
1153 odsPrinter << stringifyImageDepthInfo(getValue());
1154 odsPrinter << ">";
1155}
1156
1157::mlir::spirv::ImageDepthInfo ImageDepthInfoAttr::getValue() const {
1158 return getImpl()->value;
1159}
1160
1161} // namespace spirv
1162} // namespace mlir
1163MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ImageDepthInfoAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ImageDepthInfoAttr>::id = {}; } }
1164namespace mlir {
1165namespace spirv {
1166namespace detail {
1167struct DeviceTypeAttrStorage : public ::mlir::AttributeStorage {
1168 using KeyTy = std::tuple<::mlir::spirv::DeviceType>;
1169 DeviceTypeAttrStorage(::mlir::spirv::DeviceType value) : value(value) {}
1170
1171 KeyTy getAsKey() const {
1172 return KeyTy(value);
1173 }
1174
1175 bool operator==(const KeyTy &tblgenKey) const {
1176 return (value == std::get<0>(tblgenKey));
1177 }
1178
1179 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1180 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1181 }
1182
1183 static DeviceTypeAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1184 auto value = std::get<0>(tblgenKey);
1185 return new (allocator.allocate<DeviceTypeAttrStorage>()) DeviceTypeAttrStorage(value);
1186 }
1187
1188 ::mlir::spirv::DeviceType value;
1189};
1190} // namespace detail
1191DeviceTypeAttr DeviceTypeAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::DeviceType value) {
1192 return Base::get(context, value);
1193}
1194
1195::mlir::Attribute DeviceTypeAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1196 ::mlir::Builder odsBuilder(odsParser.getContext());
1197 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1198 (void) odsLoc;
1199 ::mlir::FailureOr<::mlir::spirv::DeviceType> _result_value;
1200 // Parse literal '<'
1201 if (odsParser.parseLess()) return {};
1202
1203 // Parse variable 'value'
1204 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::DeviceType> {
1205 auto loc = odsParser.getCurrentLocation();
1206 ::llvm::StringRef enumKeyword;
1207 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1208 return ::mlir::failure();
1209 auto maybeEnum = ::mlir::spirv::symbolizeDeviceType(enumKeyword);
1210 if (maybeEnum)
1211 return *maybeEnum;
1212 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::DeviceType" << " to be one of: " << "Other" << ", " << "IntegratedGPU" << ", " << "DiscreteGPU" << ", " << "CPU" << ", " << "Unknown")};
1213 }();
1214 if (::mlir::failed(_result_value)) {
1215 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_DeviceTypeAttr parameter 'value' which is to be a `::mlir::spirv::DeviceType`");
1216 return {};
1217 }
1218 // Parse literal '>'
1219 if (odsParser.parseGreater()) return {};
1220 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"
, 1220, __extension__ __PRETTY_FUNCTION__));
1221 return DeviceTypeAttr::get(odsParser.getContext(),
1222 ::mlir::spirv::DeviceType((*_result_value)));
1223}
1224
1225void DeviceTypeAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1226 ::mlir::Builder odsBuilder(getContext());
1227 odsPrinter << "<";
1228 odsPrinter << stringifyDeviceType(getValue());
1229 odsPrinter << ">";
1230}
1231
1232::mlir::spirv::DeviceType DeviceTypeAttr::getValue() const {
1233 return getImpl()->value;
1234}
1235
1236} // namespace spirv
1237} // namespace mlir
1238MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::DeviceTypeAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::DeviceTypeAttr>::id = {}; } }
1239namespace mlir {
1240namespace spirv {
1241namespace detail {
1242struct DimAttrStorage : public ::mlir::AttributeStorage {
1243 using KeyTy = std::tuple<::mlir::spirv::Dim>;
1244 DimAttrStorage(::mlir::spirv::Dim value) : value(value) {}
1245
1246 KeyTy getAsKey() const {
1247 return KeyTy(value);
1248 }
1249
1250 bool operator==(const KeyTy &tblgenKey) const {
1251 return (value == std::get<0>(tblgenKey));
1252 }
1253
1254 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1255 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1256 }
1257
1258 static DimAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1259 auto value = std::get<0>(tblgenKey);
1260 return new (allocator.allocate<DimAttrStorage>()) DimAttrStorage(value);
1261 }
1262
1263 ::mlir::spirv::Dim value;
1264};
1265} // namespace detail
1266DimAttr DimAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Dim value) {
1267 return Base::get(context, value);
1268}
1269
1270::mlir::Attribute DimAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1271 ::mlir::Builder odsBuilder(odsParser.getContext());
1272 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1273 (void) odsLoc;
1274 ::mlir::FailureOr<::mlir::spirv::Dim> _result_value;
1275 // Parse literal '<'
1276 if (odsParser.parseLess()) return {};
1277
1278 // Parse variable 'value'
1279 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Dim> {
1280 auto loc = odsParser.getCurrentLocation();
1281 ::llvm::StringRef enumKeyword;
1282 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1283 return ::mlir::failure();
1284 auto maybeEnum = ::mlir::spirv::symbolizeDim(enumKeyword);
1285 if (maybeEnum)
1286 return *maybeEnum;
1287 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::Dim" << " to be one of: " << "Dim1D" << ", " << "Dim2D" << ", " << "Dim3D" << ", " << "Cube" << ", " << "Rect" << ", " << "Buffer" << ", " << "SubpassData")};
1288 }();
1289 if (::mlir::failed(_result_value)) {
1290 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_DimAttr parameter 'value' which is to be a `::mlir::spirv::Dim`");
1291 return {};
1292 }
1293 // Parse literal '>'
1294 if (odsParser.parseGreater()) return {};
1295 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"
, 1295, __extension__ __PRETTY_FUNCTION__));
1296 return DimAttr::get(odsParser.getContext(),
1297 ::mlir::spirv::Dim((*_result_value)));
1298}
1299
1300void DimAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1301 ::mlir::Builder odsBuilder(getContext());
1302 odsPrinter << "<";
1303 odsPrinter << stringifyDim(getValue());
1304 odsPrinter << ">";
1305}
1306
1307::mlir::spirv::Dim DimAttr::getValue() const {
1308 return getImpl()->value;
1309}
1310
1311} // namespace spirv
1312} // namespace mlir
1313MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::DimAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::DimAttr>::id = {}; } }
1314namespace mlir {
1315namespace spirv {
1316namespace detail {
1317struct EntryPointABIAttrStorage : public ::mlir::AttributeStorage {
1318 using KeyTy = std::tuple<DenseI32ArrayAttr, std::optional<int>>;
1319 EntryPointABIAttrStorage(DenseI32ArrayAttr workgroup_size, std::optional<int> subgroup_size) : workgroup_size(workgroup_size), subgroup_size(subgroup_size) {}
1320
1321 KeyTy getAsKey() const {
1322 return KeyTy(workgroup_size, subgroup_size);
1323 }
1324
1325 bool operator==(const KeyTy &tblgenKey) const {
1326 return (workgroup_size == std::get<0>(tblgenKey)) && (subgroup_size == std::get<1>(tblgenKey));
1327 }
1328
1329 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1330 return ::llvm::hash_combine(std::get<0>(tblgenKey), std::get<1>(tblgenKey));
1331 }
1332
1333 static EntryPointABIAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1334 auto workgroup_size = std::get<0>(tblgenKey);
1335 auto subgroup_size = std::get<1>(tblgenKey);
1336 return new (allocator.allocate<EntryPointABIAttrStorage>()) EntryPointABIAttrStorage(workgroup_size, subgroup_size);
1337 }
1338
1339 DenseI32ArrayAttr workgroup_size;
1340 std::optional<int> subgroup_size;
1341};
1342} // namespace detail
1343EntryPointABIAttr EntryPointABIAttr::get(::mlir::MLIRContext *context, DenseI32ArrayAttr workgroup_size, std::optional<int> subgroup_size) {
1344 return Base::get(context, workgroup_size, subgroup_size);
1345}
1346
1347::mlir::Attribute EntryPointABIAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1348 ::mlir::Builder odsBuilder(odsParser.getContext());
1349 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1350 (void) odsLoc;
1351 ::mlir::FailureOr<DenseI32ArrayAttr> _result_workgroup_size;
1352 ::mlir::FailureOr<std::optional<int>> _result_subgroup_size;
1353 // Parse literal '<'
1354 if (odsParser.parseLess()) return {};
1355 // Parse parameter struct
1356 bool _seen_workgroup_size = false;
1357 bool _seen_subgroup_size = false;
1358 {
1359 const auto _loop_body = [&](::llvm::StringRef _paramKey) -> bool {
1360 // Parse literal '='
1361 if (odsParser.parseEqual()) return {};
1362 if (!_seen_workgroup_size && _paramKey == "workgroup_size") {
1363 _seen_workgroup_size = true;
1364
1365 // Parse variable 'workgroup_size'
1366 _result_workgroup_size = ::mlir::FieldParser<DenseI32ArrayAttr>::parse(odsParser);
1367 if (::mlir::failed(_result_workgroup_size)) {
1368 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_EntryPointABIAttr parameter 'workgroup_size' which is to be a `DenseI32ArrayAttr`");
1369 return {};
1370 }
1371 } else if (!_seen_subgroup_size && _paramKey == "subgroup_size") {
1372 _seen_subgroup_size = true;
1373
1374 // Parse variable 'subgroup_size'
1375 _result_subgroup_size = ::mlir::FieldParser<std::optional<int>>::parse(odsParser);
1376 if (::mlir::failed(_result_subgroup_size)) {
1377 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_EntryPointABIAttr parameter 'subgroup_size' which is to be a `std::optional<int>`");
1378 return {};
1379 }
1380 } else {
1381 odsParser.emitError(odsParser.getCurrentLocation(), "duplicate or unknown struct parameter name: ") << _paramKey;
1382 return {};
1383 }
1384 return true;
1385 };
1386 ::llvm::StringRef _paramKey;
1387 if (!odsParser.parseOptionalKeyword(&_paramKey)) {
1388 if (!_loop_body(_paramKey)) return {};
1389 while (!odsParser.parseOptionalComma()) {
1390 ::llvm::StringRef _paramKey;
1391 if (odsParser.parseKeyword(&_paramKey)) {
1392 odsParser.emitError(odsParser.getCurrentLocation(),
1393 "expected a parameter name in struct");
1394 return {};
1395 }
1396 if (!_loop_body(_paramKey)) return {};
1397 }
1398 }
1399 }
1400 // Parse literal '>'
1401 if (odsParser.parseGreater()) return {};
1402 return EntryPointABIAttr::get(odsParser.getContext(),
1403 DenseI32ArrayAttr((_result_workgroup_size.value_or(DenseI32ArrayAttr()))),
1404 std::optional<int>((_result_subgroup_size.value_or(std::optional<int>()))));
1405}
1406
1407void EntryPointABIAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1408 ::mlir::Builder odsBuilder(getContext());
1409 odsPrinter << "<";
1410 {
1411 bool _firstPrinted = true;
1412 if (!(getWorkgroupSize() == DenseI32ArrayAttr())) {
1413 if (!_firstPrinted) odsPrinter << ", ";
1414 _firstPrinted = false;
1415 odsPrinter << "workgroup_size = ";
1416 if (!(getWorkgroupSize() == DenseI32ArrayAttr())) {
1417 odsPrinter.printStrippedAttrOrType(getWorkgroupSize());
1418 }
1419 }
1420 if (!(getSubgroupSize() == std::optional<int>())) {
1421 if (!_firstPrinted) odsPrinter << ", ";
1422 _firstPrinted = false;
1423 odsPrinter << "subgroup_size = ";
1424 if (!(getSubgroupSize() == std::optional<int>())) {
1425 odsPrinter.printStrippedAttrOrType(getSubgroupSize());
1426 }
1427 }
1428 }
1429 odsPrinter << ">";
1430}
1431
1432DenseI32ArrayAttr EntryPointABIAttr::getWorkgroupSize() const {
1433 return getImpl()->workgroup_size;
1434}
1435
1436std::optional<int> EntryPointABIAttr::getSubgroupSize() const {
1437 return getImpl()->subgroup_size;
1438}
1439
1440} // namespace spirv
1441} // namespace mlir
1442MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::EntryPointABIAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::EntryPointABIAttr>::id = {}; } }
1443namespace mlir {
1444namespace spirv {
1445namespace detail {
1446struct ExecutionModeAttrStorage : public ::mlir::AttributeStorage {
1447 using KeyTy = std::tuple<::mlir::spirv::ExecutionMode>;
1448 ExecutionModeAttrStorage(::mlir::spirv::ExecutionMode value) : value(value) {}
1449
1450 KeyTy getAsKey() const {
1451 return KeyTy(value);
1452 }
1453
1454 bool operator==(const KeyTy &tblgenKey) const {
1455 return (value == std::get<0>(tblgenKey));
1456 }
1457
1458 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1459 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1460 }
1461
1462 static ExecutionModeAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1463 auto value = std::get<0>(tblgenKey);
1464 return new (allocator.allocate<ExecutionModeAttrStorage>()) ExecutionModeAttrStorage(value);
1465 }
1466
1467 ::mlir::spirv::ExecutionMode value;
1468};
1469} // namespace detail
1470ExecutionModeAttr ExecutionModeAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ExecutionMode value) {
1471 return Base::get(context, value);
1472}
1473
1474::mlir::Attribute ExecutionModeAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1475 ::mlir::Builder odsBuilder(odsParser.getContext());
1476 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1477 (void) odsLoc;
1478 ::mlir::FailureOr<::mlir::spirv::ExecutionMode> _result_value;
1479 // Parse literal '<'
1480 if (odsParser.parseLess()) return {};
1481
1482 // Parse variable 'value'
1483 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ExecutionMode> {
1484 auto loc = odsParser.getCurrentLocation();
1485 ::llvm::StringRef enumKeyword;
1486 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1487 return ::mlir::failure();
1488 auto maybeEnum = ::mlir::spirv::symbolizeExecutionMode(enumKeyword);
1489 if (maybeEnum)
1490 return *maybeEnum;
1491 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" << ", " << "StreamingInterfaceINTEL" << ", " << "NamedBarrierCountINTEL")};
1492 }();
1493 if (::mlir::failed(_result_value)) {
1494 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ExecutionModeAttr parameter 'value' which is to be a `::mlir::spirv::ExecutionMode`");
1495 return {};
1496 }
1497 // Parse literal '>'
1498 if (odsParser.parseGreater()) return {};
1499 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"
, 1499, __extension__ __PRETTY_FUNCTION__));
1500 return ExecutionModeAttr::get(odsParser.getContext(),
1501 ::mlir::spirv::ExecutionMode((*_result_value)));
1502}
1503
1504void ExecutionModeAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1505 ::mlir::Builder odsBuilder(getContext());
1506 odsPrinter << "<";
1507 odsPrinter << stringifyExecutionMode(getValue());
1508 odsPrinter << ">";
1509}
1510
1511::mlir::spirv::ExecutionMode ExecutionModeAttr::getValue() const {
1512 return getImpl()->value;
1513}
1514
1515} // namespace spirv
1516} // namespace mlir
1517MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ExecutionModeAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ExecutionModeAttr>::id = {}; } }
1518namespace mlir {
1519namespace spirv {
1520namespace detail {
1521struct ExecutionModelAttrStorage : public ::mlir::AttributeStorage {
1522 using KeyTy = std::tuple<::mlir::spirv::ExecutionModel>;
1523 ExecutionModelAttrStorage(::mlir::spirv::ExecutionModel value) : value(value) {}
1524
1525 KeyTy getAsKey() const {
1526 return KeyTy(value);
1527 }
1528
1529 bool operator==(const KeyTy &tblgenKey) const {
1530 return (value == std::get<0>(tblgenKey));
1531 }
1532
1533 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1534 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1535 }
1536
1537 static ExecutionModelAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1538 auto value = std::get<0>(tblgenKey);
1539 return new (allocator.allocate<ExecutionModelAttrStorage>()) ExecutionModelAttrStorage(value);
1540 }
1541
1542 ::mlir::spirv::ExecutionModel value;
1543};
1544} // namespace detail
1545ExecutionModelAttr ExecutionModelAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ExecutionModel value) {
1546 return Base::get(context, value);
1547}
1548
1549::mlir::Attribute ExecutionModelAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1550 ::mlir::Builder odsBuilder(odsParser.getContext());
1551 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1552 (void) odsLoc;
1553 ::mlir::FailureOr<::mlir::spirv::ExecutionModel> _result_value;
1554 // Parse literal '<'
1555 if (odsParser.parseLess()) return {};
1556
1557 // Parse variable 'value'
1558 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ExecutionModel> {
1559 auto loc = odsParser.getCurrentLocation();
1560 ::llvm::StringRef enumKeyword;
1561 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1562 return ::mlir::failure();
1563 auto maybeEnum = ::mlir::spirv::symbolizeExecutionModel(enumKeyword);
1564 if (maybeEnum)
1565 return *maybeEnum;
1566 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")};
1567 }();
1568 if (::mlir::failed(_result_value)) {
1569 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ExecutionModelAttr parameter 'value' which is to be a `::mlir::spirv::ExecutionModel`");
1570 return {};
1571 }
1572 // Parse literal '>'
1573 if (odsParser.parseGreater()) return {};
1574 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"
, 1574, __extension__ __PRETTY_FUNCTION__));
1575 return ExecutionModelAttr::get(odsParser.getContext(),
1576 ::mlir::spirv::ExecutionModel((*_result_value)));
1577}
1578
1579void ExecutionModelAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1580 ::mlir::Builder odsBuilder(getContext());
1581 odsPrinter << "<";
1582 odsPrinter << stringifyExecutionModel(getValue());
1583 odsPrinter << ">";
1584}
1585
1586::mlir::spirv::ExecutionModel ExecutionModelAttr::getValue() const {
1587 return getImpl()->value;
1588}
1589
1590} // namespace spirv
1591} // namespace mlir
1592MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ExecutionModelAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ExecutionModelAttr>::id = {}; } }
1593namespace mlir {
1594namespace spirv {
1595namespace detail {
1596struct ExtensionAttrStorage : public ::mlir::AttributeStorage {
1597 using KeyTy = std::tuple<::mlir::spirv::Extension>;
1598 ExtensionAttrStorage(::mlir::spirv::Extension value) : value(value) {}
1599
1600 KeyTy getAsKey() const {
1601 return KeyTy(value);
1602 }
1603
1604 bool operator==(const KeyTy &tblgenKey) const {
1605 return (value == std::get<0>(tblgenKey));
1606 }
1607
1608 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1609 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1610 }
1611
1612 static ExtensionAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1613 auto value = std::get<0>(tblgenKey);
1614 return new (allocator.allocate<ExtensionAttrStorage>()) ExtensionAttrStorage(value);
1615 }
1616
1617 ::mlir::spirv::Extension value;
1618};
1619} // namespace detail
1620ExtensionAttr ExtensionAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Extension value) {
1621 return Base::get(context, value);
1622}
1623
1624::mlir::Attribute ExtensionAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1625 ::mlir::Builder odsBuilder(odsParser.getContext());
1626 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1627 (void) odsLoc;
1628 ::mlir::FailureOr<::mlir::spirv::Extension> _result_value;
1629 // Parse literal '<'
1630 if (odsParser.parseLess()) return {};
1631
1632 // Parse variable 'value'
1633 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Extension> {
1634 auto loc = odsParser.getCurrentLocation();
1635 ::llvm::StringRef enumKeyword;
1636 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1637 return ::mlir::failure();
1638 auto maybeEnum = ::mlir::spirv::symbolizeExtension(enumKeyword);
1639 if (maybeEnum)
1640 return *maybeEnum;
1641 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_KHR_non_semantic_info" << ", " << "SPV_KHR_terminate_invocation" << ", " << "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_INTEL_bfloat16_conversion" << ", " << "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")};
1642 }();
1643 if (::mlir::failed(_result_value)) {
1644 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ExtensionAttr parameter 'value' which is to be a `::mlir::spirv::Extension`");
1645 return {};
1646 }
1647 // Parse literal '>'
1648 if (odsParser.parseGreater()) return {};
1649 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"
, 1649, __extension__ __PRETTY_FUNCTION__));
1650 return ExtensionAttr::get(odsParser.getContext(),
1651 ::mlir::spirv::Extension((*_result_value)));
1652}
1653
1654void ExtensionAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1655 ::mlir::Builder odsBuilder(getContext());
1656 odsPrinter << "<";
1657 odsPrinter << stringifyExtension(getValue());
1658 odsPrinter << ">";
1659}
1660
1661::mlir::spirv::Extension ExtensionAttr::getValue() const {
1662 return getImpl()->value;
1663}
1664
1665} // namespace spirv
1666} // namespace mlir
1667MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ExtensionAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ExtensionAttr>::id = {}; } }
1668namespace mlir {
1669namespace spirv {
1670namespace detail {
1671struct FunctionControlAttrStorage : public ::mlir::AttributeStorage {
1672 using KeyTy = std::tuple<::mlir::spirv::FunctionControl>;
1673 FunctionControlAttrStorage(::mlir::spirv::FunctionControl value) : value(value) {}
1674
1675 KeyTy getAsKey() const {
1676 return KeyTy(value);
1677 }
1678
1679 bool operator==(const KeyTy &tblgenKey) const {
1680 return (value == std::get<0>(tblgenKey));
1681 }
1682
1683 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1684 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1685 }
1686
1687 static FunctionControlAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1688 auto value = std::get<0>(tblgenKey);
1689 return new (allocator.allocate<FunctionControlAttrStorage>()) FunctionControlAttrStorage(value);
1690 }
1691
1692 ::mlir::spirv::FunctionControl value;
1693};
1694} // namespace detail
1695FunctionControlAttr FunctionControlAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::FunctionControl value) {
1696 return Base::get(context, value);
1697}
1698
1699::mlir::Attribute FunctionControlAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1700 ::mlir::Builder odsBuilder(odsParser.getContext());
1701 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1702 (void) odsLoc;
1703 ::mlir::FailureOr<::mlir::spirv::FunctionControl> _result_value;
1704 // Parse literal '<'
1705 if (odsParser.parseLess()) return {};
1706
1707 // Parse variable 'value'
1708 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::FunctionControl> {
1709 ::mlir::spirv::FunctionControl flags = {};
1710 auto loc = odsParser.getCurrentLocation();
1711 ::llvm::StringRef enumKeyword;
1712 do {
1713 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1714 return ::mlir::failure();
1715 auto maybeEnum = ::mlir::spirv::symbolizeFunctionControl(enumKeyword);
1716 if (!maybeEnum) {
1717 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::FunctionControl" << " to be one of: " << "None" << ", " << "Inline" << ", " << "DontInline" << ", " << "Pure" << ", " << "Const" << ", " << "OptNoneINTEL")};
1718 }
1719 flags = flags | *maybeEnum;
1720 } while(::mlir::succeeded(odsParser.parseOptionalVerticalBar()));
1721 return flags;
1722 }();
1723 if (::mlir::failed(_result_value)) {
1724 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_FunctionControlAttr parameter 'value' which is to be a `::mlir::spirv::FunctionControl`");
1725 return {};
1726 }
1727 // Parse literal '>'
1728 if (odsParser.parseGreater()) return {};
1729 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"
, 1729, __extension__ __PRETTY_FUNCTION__));
1730 return FunctionControlAttr::get(odsParser.getContext(),
1731 ::mlir::spirv::FunctionControl((*_result_value)));
1732}
1733
1734void FunctionControlAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1735 ::mlir::Builder odsBuilder(getContext());
1736 odsPrinter << "<";
1737 odsPrinter << stringifyFunctionControl(getValue());
1738 odsPrinter << ">";
1739}
1740
1741::mlir::spirv::FunctionControl FunctionControlAttr::getValue() const {
1742 return getImpl()->value;
1743}
1744
1745} // namespace spirv
1746} // namespace mlir
1747MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::FunctionControlAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::FunctionControlAttr>::id = {}; } }
1748namespace mlir {
1749namespace spirv {
1750namespace detail {
1751struct GroupOperationAttrStorage : public ::mlir::AttributeStorage {
1752 using KeyTy = std::tuple<::mlir::spirv::GroupOperation>;
1753 GroupOperationAttrStorage(::mlir::spirv::GroupOperation value) : value(value) {}
1754
1755 KeyTy getAsKey() const {
1756 return KeyTy(value);
1757 }
1758
1759 bool operator==(const KeyTy &tblgenKey) const {
1760 return (value == std::get<0>(tblgenKey));
1761 }
1762
1763 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1764 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1765 }
1766
1767 static GroupOperationAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1768 auto value = std::get<0>(tblgenKey);
1769 return new (allocator.allocate<GroupOperationAttrStorage>()) GroupOperationAttrStorage(value);
1770 }
1771
1772 ::mlir::spirv::GroupOperation value;
1773};
1774} // namespace detail
1775GroupOperationAttr GroupOperationAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::GroupOperation value) {
1776 return Base::get(context, value);
1777}
1778
1779::mlir::Attribute GroupOperationAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1780 ::mlir::Builder odsBuilder(odsParser.getContext());
1781 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1782 (void) odsLoc;
1783 ::mlir::FailureOr<::mlir::spirv::GroupOperation> _result_value;
1784 // Parse literal '<'
1785 if (odsParser.parseLess()) return {};
1786
1787 // Parse variable 'value'
1788 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::GroupOperation> {
1789 auto loc = odsParser.getCurrentLocation();
1790 ::llvm::StringRef enumKeyword;
1791 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1792 return ::mlir::failure();
1793 auto maybeEnum = ::mlir::spirv::symbolizeGroupOperation(enumKeyword);
1794 if (maybeEnum)
1795 return *maybeEnum;
1796 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::GroupOperation" << " to be one of: " << "Reduce" << ", " << "InclusiveScan" << ", " << "ExclusiveScan" << ", " << "ClusteredReduce" << ", " << "PartitionedReduceNV" << ", " << "PartitionedInclusiveScanNV" << ", " << "PartitionedExclusiveScanNV")};
1797 }();
1798 if (::mlir::failed(_result_value)) {
1799 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_GroupOperationAttr parameter 'value' which is to be a `::mlir::spirv::GroupOperation`");
1800 return {};
1801 }
1802 // Parse literal '>'
1803 if (odsParser.parseGreater()) return {};
1804 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"
, 1804, __extension__ __PRETTY_FUNCTION__));
1805 return GroupOperationAttr::get(odsParser.getContext(),
1806 ::mlir::spirv::GroupOperation((*_result_value)));
1807}
1808
1809void GroupOperationAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1810 ::mlir::Builder odsBuilder(getContext());
1811 odsPrinter << "<";
1812 odsPrinter << stringifyGroupOperation(getValue());
1813 odsPrinter << ">";
1814}
1815
1816::mlir::spirv::GroupOperation GroupOperationAttr::getValue() const {
1817 return getImpl()->value;
1818}
1819
1820} // namespace spirv
1821} // namespace mlir
1822MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::GroupOperationAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::GroupOperationAttr>::id = {}; } }
1823namespace mlir {
1824namespace spirv {
1825namespace detail {
1826struct ImageFormatAttrStorage : public ::mlir::AttributeStorage {
1827 using KeyTy = std::tuple<::mlir::spirv::ImageFormat>;
1828 ImageFormatAttrStorage(::mlir::spirv::ImageFormat value) : value(value) {}
1829
1830 KeyTy getAsKey() const {
1831 return KeyTy(value);
1832 }
1833
1834 bool operator==(const KeyTy &tblgenKey) const {
1835 return (value == std::get<0>(tblgenKey));
1836 }
1837
1838 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1839 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1840 }
1841
1842 static ImageFormatAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1843 auto value = std::get<0>(tblgenKey);
1844 return new (allocator.allocate<ImageFormatAttrStorage>()) ImageFormatAttrStorage(value);
1845 }
1846
1847 ::mlir::spirv::ImageFormat value;
1848};
1849} // namespace detail
1850ImageFormatAttr ImageFormatAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ImageFormat value) {
1851 return Base::get(context, value);
1852}
1853
1854::mlir::Attribute ImageFormatAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1855 ::mlir::Builder odsBuilder(odsParser.getContext());
1856 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1857 (void) odsLoc;
1858 ::mlir::FailureOr<::mlir::spirv::ImageFormat> _result_value;
1859 // Parse literal '<'
1860 if (odsParser.parseLess()) return {};
1861
1862 // Parse variable 'value'
1863 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ImageFormat> {
1864 auto loc = odsParser.getCurrentLocation();
1865 ::llvm::StringRef enumKeyword;
1866 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1867 return ::mlir::failure();
1868 auto maybeEnum = ::mlir::spirv::symbolizeImageFormat(enumKeyword);
1869 if (maybeEnum)
1870 return *maybeEnum;
1871 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")};
1872 }();
1873 if (::mlir::failed(_result_value)) {
1874 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ImageFormatAttr parameter 'value' which is to be a `::mlir::spirv::ImageFormat`");
1875 return {};
1876 }
1877 // Parse literal '>'
1878 if (odsParser.parseGreater()) return {};
1879 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"
, 1879, __extension__ __PRETTY_FUNCTION__));
1880 return ImageFormatAttr::get(odsParser.getContext(),
1881 ::mlir::spirv::ImageFormat((*_result_value)));
1882}
1883
1884void ImageFormatAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1885 ::mlir::Builder odsBuilder(getContext());
1886 odsPrinter << "<";
1887 odsPrinter << stringifyImageFormat(getValue());
1888 odsPrinter << ">";
1889}
1890
1891::mlir::spirv::ImageFormat ImageFormatAttr::getValue() const {
1892 return getImpl()->value;
1893}
1894
1895} // namespace spirv
1896} // namespace mlir
1897MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ImageFormatAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ImageFormatAttr>::id = {}; } }
1898namespace mlir {
1899namespace spirv {
1900namespace detail {
1901struct ImageOperandsAttrStorage : public ::mlir::AttributeStorage {
1902 using KeyTy = std::tuple<::mlir::spirv::ImageOperands>;
1903 ImageOperandsAttrStorage(::mlir::spirv::ImageOperands value) : value(value) {}
1904
1905 KeyTy getAsKey() const {
1906 return KeyTy(value);
1907 }
1908
1909 bool operator==(const KeyTy &tblgenKey) const {
1910 return (value == std::get<0>(tblgenKey));
1911 }
1912
1913 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1914 return ::llvm::hash_combine(std::get<0>(tblgenKey));
1915 }
1916
1917 static ImageOperandsAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1918 auto value = std::get<0>(tblgenKey);
1919 return new (allocator.allocate<ImageOperandsAttrStorage>()) ImageOperandsAttrStorage(value);
1920 }
1921
1922 ::mlir::spirv::ImageOperands value;
1923};
1924} // namespace detail
1925ImageOperandsAttr ImageOperandsAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ImageOperands value) {
1926 return Base::get(context, value);
1927}
1928
1929::mlir::Attribute ImageOperandsAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
1930 ::mlir::Builder odsBuilder(odsParser.getContext());
1931 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
1932 (void) odsLoc;
1933 ::mlir::FailureOr<::mlir::spirv::ImageOperands> _result_value;
1934 // Parse literal '<'
1935 if (odsParser.parseLess()) return {};
1936
1937 // Parse variable 'value'
1938 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ImageOperands> {
1939 ::mlir::spirv::ImageOperands flags = {};
1940 auto loc = odsParser.getCurrentLocation();
1941 ::llvm::StringRef enumKeyword;
1942 do {
1943 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
1944 return ::mlir::failure();
1945 auto maybeEnum = ::mlir::spirv::symbolizeImageOperands(enumKeyword);
1946 if (!maybeEnum) {
1947 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")};
1948 }
1949 flags = flags | *maybeEnum;
1950 } while(::mlir::succeeded(odsParser.parseOptionalVerticalBar()));
1951 return flags;
1952 }();
1953 if (::mlir::failed(_result_value)) {
1954 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ImageOperandsAttr parameter 'value' which is to be a `::mlir::spirv::ImageOperands`");
1955 return {};
1956 }
1957 // Parse literal '>'
1958 if (odsParser.parseGreater()) return {};
1959 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"
, 1959, __extension__ __PRETTY_FUNCTION__));
1960 return ImageOperandsAttr::get(odsParser.getContext(),
1961 ::mlir::spirv::ImageOperands((*_result_value)));
1962}
1963
1964void ImageOperandsAttr::print(::mlir::AsmPrinter &odsPrinter) const {
1965 ::mlir::Builder odsBuilder(getContext());
1966 odsPrinter << "<";
1967 odsPrinter << stringifyImageOperands(getValue());
1968 odsPrinter << ">";
1969}
1970
1971::mlir::spirv::ImageOperands ImageOperandsAttr::getValue() const {
1972 return getImpl()->value;
1973}
1974
1975} // namespace spirv
1976} // namespace mlir
1977MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ImageOperandsAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ImageOperandsAttr>::id = {}; } }
1978namespace mlir {
1979namespace spirv {
1980namespace detail {
1981struct JointMatrixPropertiesINTELAttrStorage : public ::mlir::AttributeStorage {
1982 using KeyTy = std::tuple<int, int, int, mlir::Type, mlir::Type, mlir::Type, mlir::Type, mlir::spirv::ScopeAttr>;
1983 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) {}
1984
1985 KeyTy getAsKey() const {
1986 return KeyTy(m_size, n_size, k_size, a_type, b_type, c_type, result_type, scope);
1987 }
1988
1989 bool operator==(const KeyTy &tblgenKey) const {
1990 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));
1991 }
1992
1993 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
1994 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));
1995 }
1996
1997 static JointMatrixPropertiesINTELAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
1998 auto m_size = std::get<0>(tblgenKey);
1999 auto n_size = std::get<1>(tblgenKey);
2000 auto k_size = std::get<2>(tblgenKey);
2001 auto a_type = std::get<3>(tblgenKey);
2002 auto b_type = std::get<4>(tblgenKey);
2003 auto c_type = std::get<5>(tblgenKey);
2004 auto result_type = std::get<6>(tblgenKey);
2005 auto scope = std::get<7>(tblgenKey);
2006 return new (allocator.allocate<JointMatrixPropertiesINTELAttrStorage>()) JointMatrixPropertiesINTELAttrStorage(m_size, n_size, k_size, a_type, b_type, c_type, result_type, scope);
2007 }
2008
2009 int m_size;
2010 int n_size;
2011 int k_size;
2012 mlir::Type a_type;
2013 mlir::Type b_type;
2014 mlir::Type c_type;
2015 mlir::Type result_type;
2016 mlir::spirv::ScopeAttr scope;
2017};
2018} // namespace detail
2019JointMatrixPropertiesINTELAttr 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) {
2020 return Base::get(context, m_size, n_size, k_size, a_type, b_type, c_type, result_type, scope);
2021}
2022
2023::mlir::Attribute JointMatrixPropertiesINTELAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2024 ::mlir::Builder odsBuilder(odsParser.getContext());
2025 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2026 (void) odsLoc;
2027 ::mlir::FailureOr<int> _result_m_size;
2028 ::mlir::FailureOr<int> _result_n_size;
2029 ::mlir::FailureOr<int> _result_k_size;
2030 ::mlir::FailureOr<mlir::Type> _result_a_type;
2031 ::mlir::FailureOr<mlir::Type> _result_b_type;
2032 ::mlir::FailureOr<mlir::Type> _result_c_type;
2033 ::mlir::FailureOr<mlir::Type> _result_result_type;
2034 ::mlir::FailureOr<mlir::spirv::ScopeAttr> _result_scope;
2035 // Parse literal '<'
2036 if (odsParser.parseLess()) return {};
2037 // Parse parameter struct
2038 bool _seen_m_size = false;
2039 bool _seen_n_size = false;
2040 bool _seen_k_size = false;
2041 bool _seen_a_type = false;
2042 bool _seen_b_type = false;
2043 bool _seen_c_type = false;
2044 bool _seen_result_type = false;
2045 bool _seen_scope = false;
2046 {
2047 const auto _loop_body = [&](::llvm::StringRef _paramKey) -> bool {
2048 // Parse literal '='
2049 if (odsParser.parseEqual()) return {};
2050 if (!_seen_m_size && _paramKey == "m_size") {
2051 _seen_m_size = true;
2052
2053 // Parse variable 'm_size'
2054 _result_m_size = ::mlir::FieldParser<int>::parse(odsParser);
2055 if (::mlir::failed(_result_m_size)) {
2056 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'm_size' which is to be a `int`");
2057 return {};
2058 }
2059 } else if (!_seen_n_size && _paramKey == "n_size") {
2060 _seen_n_size = true;
2061
2062 // Parse variable 'n_size'
2063 _result_n_size = ::mlir::FieldParser<int>::parse(odsParser);
2064 if (::mlir::failed(_result_n_size)) {
2065 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'n_size' which is to be a `int`");
2066 return {};
2067 }
2068 } else if (!_seen_k_size && _paramKey == "k_size") {
2069 _seen_k_size = true;
2070
2071 // Parse variable 'k_size'
2072 _result_k_size = ::mlir::FieldParser<int>::parse(odsParser);
2073 if (::mlir::failed(_result_k_size)) {
2074 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'k_size' which is to be a `int`");
2075 return {};
2076 }
2077 } else if (!_seen_a_type && _paramKey == "a_type") {
2078 _seen_a_type = true;
2079
2080 // Parse variable 'a_type'
2081 _result_a_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
2082 if (::mlir::failed(_result_a_type)) {
2083 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'a_type' which is to be a `mlir::Type`");
2084 return {};
2085 }
2086 } else if (!_seen_b_type && _paramKey == "b_type") {
2087 _seen_b_type = true;
2088
2089 // Parse variable 'b_type'
2090 _result_b_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
2091 if (::mlir::failed(_result_b_type)) {
2092 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'b_type' which is to be a `mlir::Type`");
2093 return {};
2094 }
2095 } else if (!_seen_c_type && _paramKey == "c_type") {
2096 _seen_c_type = true;
2097
2098 // Parse variable 'c_type'
2099 _result_c_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
2100 if (::mlir::failed(_result_c_type)) {
2101 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'c_type' which is to be a `mlir::Type`");
2102 return {};
2103 }
2104 } else if (!_seen_result_type && _paramKey == "result_type") {
2105 _seen_result_type = true;
2106
2107 // Parse variable 'result_type'
2108 _result_result_type = ::mlir::FieldParser<mlir::Type>::parse(odsParser);
2109 if (::mlir::failed(_result_result_type)) {
2110 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'result_type' which is to be a `mlir::Type`");
2111 return {};
2112 }
2113 } else if (!_seen_scope && _paramKey == "scope") {
2114 _seen_scope = true;
2115
2116 // Parse variable 'scope'
2117 _result_scope = ::mlir::FieldParser<mlir::spirv::ScopeAttr>::parse(odsParser);
2118 if (::mlir::failed(_result_scope)) {
2119 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_JointMatrixPropertiesINTELAttr parameter 'scope' which is to be a `mlir::spirv::ScopeAttr`");
2120 return {};
2121 }
2122 } else {
2123 odsParser.emitError(odsParser.getCurrentLocation(), "duplicate or unknown struct parameter name: ") << _paramKey;
2124 return {};
2125 }
2126 return true;
2127 };
2128 for (unsigned odsStructIndex = 0; odsStructIndex < 8; ++odsStructIndex) {
2129 ::llvm::StringRef _paramKey;
2130 if (odsParser.parseKeyword(&_paramKey)) {
2131 odsParser.emitError(odsParser.getCurrentLocation(),
2132 "expected a parameter name in struct");
2133 return {};
2134 }
2135 if (!_loop_body(_paramKey)) return {};
2136 if ((odsStructIndex != 8 - 1) && odsParser.parseComma())
2137 return {};
2138 }
2139 }
2140 // Parse literal '>'
2141 if (odsParser.parseGreater()) return {};
2142 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"
, 2142, __extension__ __PRETTY_FUNCTION__));
2143 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"
, 2143, __extension__ __PRETTY_FUNCTION__));
2144 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"
, 2144, __extension__ __PRETTY_FUNCTION__));
2145 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"
, 2145, __extension__ __PRETTY_FUNCTION__));
2146 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"
, 2146, __extension__ __PRETTY_FUNCTION__));
2147 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"
, 2147, __extension__ __PRETTY_FUNCTION__));
2148 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"
, 2148, __extension__ __PRETTY_FUNCTION__));
2149 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"
, 2149, __extension__ __PRETTY_FUNCTION__));
2150 return JointMatrixPropertiesINTELAttr::get(odsParser.getContext(),
2151 int((*_result_m_size)),
2152 int((*_result_n_size)),
2153 int((*_result_k_size)),
2154 mlir::Type((*_result_a_type)),
2155 mlir::Type((*_result_b_type)),
2156 mlir::Type((*_result_c_type)),
2157 mlir::Type((*_result_result_type)),
2158 mlir::spirv::ScopeAttr((*_result_scope)));
2159}
2160
2161void JointMatrixPropertiesINTELAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2162 ::mlir::Builder odsBuilder(getContext());
2163 odsPrinter << "<";
2164 {
2165 bool _firstPrinted = true;
2166 if (!_firstPrinted) odsPrinter << ", ";
2167 _firstPrinted = false;
2168 odsPrinter << "m_size = ";
2169 odsPrinter.printStrippedAttrOrType(getMSize());
2170 if (!_firstPrinted) odsPrinter << ", ";
2171 _firstPrinted = false;
2172 odsPrinter << "n_size = ";
2173 odsPrinter.printStrippedAttrOrType(getNSize());
2174 if (!_firstPrinted) odsPrinter << ", ";
2175 _firstPrinted = false;
2176 odsPrinter << "k_size = ";
2177 odsPrinter.printStrippedAttrOrType(getKSize());
2178 if (!_firstPrinted) odsPrinter << ", ";
2179 _firstPrinted = false;
2180 odsPrinter << "a_type = ";
2181 odsPrinter.printStrippedAttrOrType(getAType());
2182 if (!_firstPrinted) odsPrinter << ", ";
2183 _firstPrinted = false;
2184 odsPrinter << "b_type = ";
2185 odsPrinter.printStrippedAttrOrType(getBType());
2186 if (!_firstPrinted) odsPrinter << ", ";
2187 _firstPrinted = false;
2188 odsPrinter << "c_type = ";
2189 odsPrinter.printStrippedAttrOrType(getCType());
2190 if (!_firstPrinted) odsPrinter << ", ";
2191 _firstPrinted = false;
2192 odsPrinter << "result_type = ";
2193 odsPrinter.printStrippedAttrOrType(getResultType());
2194 if (!_firstPrinted) odsPrinter << ", ";
2195 _firstPrinted = false;
2196 odsPrinter << "scope = ";
2197 odsPrinter.printStrippedAttrOrType(getScope());
2198 }
2199 odsPrinter << ">";
2200}
2201
2202int JointMatrixPropertiesINTELAttr::getMSize() const {
2203 return getImpl()->m_size;
2204}
2205
2206int JointMatrixPropertiesINTELAttr::getNSize() const {
2207 return getImpl()->n_size;
2208}
2209
2210int JointMatrixPropertiesINTELAttr::getKSize() const {
2211 return getImpl()->k_size;
2212}
2213
2214mlir::Type JointMatrixPropertiesINTELAttr::getAType() const {
2215 return getImpl()->a_type;
2216}
2217
2218mlir::Type JointMatrixPropertiesINTELAttr::getBType() const {
2219 return getImpl()->b_type;
2220}
2221
2222mlir::Type JointMatrixPropertiesINTELAttr::getCType() const {
2223 return getImpl()->c_type;
2224}
2225
2226mlir::Type JointMatrixPropertiesINTELAttr::getResultType() const {
2227 return getImpl()->result_type;
2228}
2229
2230mlir::spirv::ScopeAttr JointMatrixPropertiesINTELAttr::getScope() const {
2231 return getImpl()->scope;
2232}
2233
2234} // namespace spirv
2235} // namespace mlir
2236MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::JointMatrixPropertiesINTELAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::JointMatrixPropertiesINTELAttr>::id = {
}; } }
2237namespace mlir {
2238namespace spirv {
2239namespace detail {
2240struct LinkageTypeAttrStorage : public ::mlir::AttributeStorage {
2241 using KeyTy = std::tuple<::mlir::spirv::LinkageType>;
2242 LinkageTypeAttrStorage(::mlir::spirv::LinkageType value) : value(value) {}
2243
2244 KeyTy getAsKey() const {
2245 return KeyTy(value);
2246 }
2247
2248 bool operator==(const KeyTy &tblgenKey) const {
2249 return (value == std::get<0>(tblgenKey));
2250 }
2251
2252 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2253 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2254 }
2255
2256 static LinkageTypeAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2257 auto value = std::get<0>(tblgenKey);
2258 return new (allocator.allocate<LinkageTypeAttrStorage>()) LinkageTypeAttrStorage(value);
2259 }
2260
2261 ::mlir::spirv::LinkageType value;
2262};
2263} // namespace detail
2264LinkageTypeAttr LinkageTypeAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::LinkageType value) {
2265 return Base::get(context, value);
2266}
2267
2268::mlir::Attribute LinkageTypeAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2269 ::mlir::Builder odsBuilder(odsParser.getContext());
2270 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2271 (void) odsLoc;
2272 ::mlir::FailureOr<::mlir::spirv::LinkageType> _result_value;
2273 // Parse literal '<'
2274 if (odsParser.parseLess()) return {};
2275
2276 // Parse variable 'value'
2277 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::LinkageType> {
2278 auto loc = odsParser.getCurrentLocation();
2279 ::llvm::StringRef enumKeyword;
2280 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2281 return ::mlir::failure();
2282 auto maybeEnum = ::mlir::spirv::symbolizeLinkageType(enumKeyword);
2283 if (maybeEnum)
2284 return *maybeEnum;
2285 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::LinkageType" << " to be one of: " << "Export" << ", " << "Import" << ", " << "LinkOnceODR")};
2286 }();
2287 if (::mlir::failed(_result_value)) {
2288 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_LinkageTypeAttr parameter 'value' which is to be a `::mlir::spirv::LinkageType`");
2289 return {};
2290 }
2291 // Parse literal '>'
2292 if (odsParser.parseGreater()) return {};
2293 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"
, 2293, __extension__ __PRETTY_FUNCTION__));
2294 return LinkageTypeAttr::get(odsParser.getContext(),
2295 ::mlir::spirv::LinkageType((*_result_value)));
2296}
2297
2298void LinkageTypeAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2299 ::mlir::Builder odsBuilder(getContext());
2300 odsPrinter << "<";
2301 odsPrinter << stringifyLinkageType(getValue());
2302 odsPrinter << ">";
2303}
2304
2305::mlir::spirv::LinkageType LinkageTypeAttr::getValue() const {
2306 return getImpl()->value;
2307}
2308
2309} // namespace spirv
2310} // namespace mlir
2311MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::LinkageTypeAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::LinkageTypeAttr>::id = {}; } }
2312namespace mlir {
2313namespace spirv {
2314namespace detail {
2315struct LoopControlAttrStorage : public ::mlir::AttributeStorage {
2316 using KeyTy = std::tuple<::mlir::spirv::LoopControl>;
2317 LoopControlAttrStorage(::mlir::spirv::LoopControl value) : value(value) {}
2318
2319 KeyTy getAsKey() const {
2320 return KeyTy(value);
2321 }
2322
2323 bool operator==(const KeyTy &tblgenKey) const {
2324 return (value == std::get<0>(tblgenKey));
2325 }
2326
2327 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2328 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2329 }
2330
2331 static LoopControlAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2332 auto value = std::get<0>(tblgenKey);
2333 return new (allocator.allocate<LoopControlAttrStorage>()) LoopControlAttrStorage(value);
2334 }
2335
2336 ::mlir::spirv::LoopControl value;
2337};
2338} // namespace detail
2339LoopControlAttr LoopControlAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::LoopControl value) {
2340 return Base::get(context, value);
2341}
2342
2343::mlir::Attribute LoopControlAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2344 ::mlir::Builder odsBuilder(odsParser.getContext());
2345 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2346 (void) odsLoc;
2347 ::mlir::FailureOr<::mlir::spirv::LoopControl> _result_value;
2348 // Parse literal '<'
2349 if (odsParser.parseLess()) return {};
2350
2351 // Parse variable 'value'
2352 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::LoopControl> {
2353 ::mlir::spirv::LoopControl flags = {};
2354 auto loc = odsParser.getCurrentLocation();
2355 ::llvm::StringRef enumKeyword;
2356 do {
2357 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2358 return ::mlir::failure();
2359 auto maybeEnum = ::mlir::spirv::symbolizeLoopControl(enumKeyword);
2360 if (!maybeEnum) {
2361 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")};
2362 }
2363 flags = flags | *maybeEnum;
2364 } while(::mlir::succeeded(odsParser.parseOptionalVerticalBar()));
2365 return flags;
2366 }();
2367 if (::mlir::failed(_result_value)) {
2368 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_LoopControlAttr parameter 'value' which is to be a `::mlir::spirv::LoopControl`");
2369 return {};
2370 }
2371 // Parse literal '>'
2372 if (odsParser.parseGreater()) return {};
2373 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"
, 2373, __extension__ __PRETTY_FUNCTION__));
2374 return LoopControlAttr::get(odsParser.getContext(),
2375 ::mlir::spirv::LoopControl((*_result_value)));
2376}
2377
2378void LoopControlAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2379 ::mlir::Builder odsBuilder(getContext());
2380 odsPrinter << "<";
2381 odsPrinter << stringifyLoopControl(getValue());
2382 odsPrinter << ">";
2383}
2384
2385::mlir::spirv::LoopControl LoopControlAttr::getValue() const {
2386 return getImpl()->value;
2387}
2388
2389} // namespace spirv
2390} // namespace mlir
2391MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::LoopControlAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::LoopControlAttr>::id = {}; } }
2392namespace mlir {
2393namespace spirv {
2394namespace detail {
2395struct MatrixLayoutAttrStorage : public ::mlir::AttributeStorage {
2396 using KeyTy = std::tuple<::mlir::spirv::MatrixLayout>;
2397 MatrixLayoutAttrStorage(::mlir::spirv::MatrixLayout value) : value(value) {}
2398
2399 KeyTy getAsKey() const {
2400 return KeyTy(value);
2401 }
2402
2403 bool operator==(const KeyTy &tblgenKey) const {
2404 return (value == std::get<0>(tblgenKey));
2405 }
2406
2407 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2408 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2409 }
2410
2411 static MatrixLayoutAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2412 auto value = std::get<0>(tblgenKey);
2413 return new (allocator.allocate<MatrixLayoutAttrStorage>()) MatrixLayoutAttrStorage(value);
2414 }
2415
2416 ::mlir::spirv::MatrixLayout value;
2417};
2418} // namespace detail
2419MatrixLayoutAttr MatrixLayoutAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::MatrixLayout value) {
2420 return Base::get(context, value);
2421}
2422
2423::mlir::Attribute MatrixLayoutAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2424 ::mlir::Builder odsBuilder(odsParser.getContext());
2425 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2426 (void) odsLoc;
2427 ::mlir::FailureOr<::mlir::spirv::MatrixLayout> _result_value;
2428 // Parse literal '<'
2429 if (odsParser.parseLess()) return {};
2430
2431 // Parse variable 'value'
2432 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::MatrixLayout> {
2433 auto loc = odsParser.getCurrentLocation();
2434 ::llvm::StringRef enumKeyword;
2435 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2436 return ::mlir::failure();
2437 auto maybeEnum = ::mlir::spirv::symbolizeMatrixLayout(enumKeyword);
2438 if (maybeEnum)
2439 return *maybeEnum;
2440 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::MatrixLayout" << " to be one of: " << "ColumnMajor" << ", " << "RowMajor" << ", " << "PackedA" << ", " << "PackedB")};
2441 }();
2442 if (::mlir::failed(_result_value)) {
2443 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_MatrixLayoutAttr parameter 'value' which is to be a `::mlir::spirv::MatrixLayout`");
2444 return {};
2445 }
2446 // Parse literal '>'
2447 if (odsParser.parseGreater()) return {};
2448 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"
, 2448, __extension__ __PRETTY_FUNCTION__));
2449 return MatrixLayoutAttr::get(odsParser.getContext(),
2450 ::mlir::spirv::MatrixLayout((*_result_value)));
2451}
2452
2453void MatrixLayoutAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2454 ::mlir::Builder odsBuilder(getContext());
2455 odsPrinter << "<";
2456 odsPrinter << stringifyMatrixLayout(getValue());
2457 odsPrinter << ">";
2458}
2459
2460::mlir::spirv::MatrixLayout MatrixLayoutAttr::getValue() const {
2461 return getImpl()->value;
2462}
2463
2464} // namespace spirv
2465} // namespace mlir
2466MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::MatrixLayoutAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::MatrixLayoutAttr>::id = {}; } }
2467namespace mlir {
2468namespace spirv {
2469namespace detail {
2470struct MemoryAccessAttrStorage : public ::mlir::AttributeStorage {
2471 using KeyTy = std::tuple<::mlir::spirv::MemoryAccess>;
2472 MemoryAccessAttrStorage(::mlir::spirv::MemoryAccess value) : value(value) {}
2473
2474 KeyTy getAsKey() const {
2475 return KeyTy(value);
2476 }
2477
2478 bool operator==(const KeyTy &tblgenKey) const {
2479 return (value == std::get<0>(tblgenKey));
2480 }
2481
2482 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2483 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2484 }
2485
2486 static MemoryAccessAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2487 auto value = std::get<0>(tblgenKey);
2488 return new (allocator.allocate<MemoryAccessAttrStorage>()) MemoryAccessAttrStorage(value);
2489 }
2490
2491 ::mlir::spirv::MemoryAccess value;
2492};
2493} // namespace detail
2494MemoryAccessAttr MemoryAccessAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::MemoryAccess value) {
2495 return Base::get(context, value);
2496}
2497
2498::mlir::Attribute MemoryAccessAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2499 ::mlir::Builder odsBuilder(odsParser.getContext());
2500 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2501 (void) odsLoc;
2502 ::mlir::FailureOr<::mlir::spirv::MemoryAccess> _result_value;
2503 // Parse literal '<'
2504 if (odsParser.parseLess()) return {};
2505
2506 // Parse variable 'value'
2507 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::MemoryAccess> {
2508 ::mlir::spirv::MemoryAccess flags = {};
2509 auto loc = odsParser.getCurrentLocation();
2510 ::llvm::StringRef enumKeyword;
2511 do {
2512 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2513 return ::mlir::failure();
2514 auto maybeEnum = ::mlir::spirv::symbolizeMemoryAccess(enumKeyword);
2515 if (!maybeEnum) {
2516 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::MemoryAccess" << " to be one of: " << "None" << ", " << "Volatile" << ", " << "Aligned" << ", " << "Nontemporal" << ", " << "MakePointerAvailable" << ", " << "MakePointerVisible" << ", " << "NonPrivatePointer" << ", " << "AliasScopeINTELMask" << ", " << "NoAliasINTELMask")};
2517 }
2518 flags = flags | *maybeEnum;
2519 } while(::mlir::succeeded(odsParser.parseOptionalVerticalBar()));
2520 return flags;
2521 }();
2522 if (::mlir::failed(_result_value)) {
2523 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_MemoryAccessAttr parameter 'value' which is to be a `::mlir::spirv::MemoryAccess`");
2524 return {};
2525 }
2526 // Parse literal '>'
2527 if (odsParser.parseGreater()) return {};
2528 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"
, 2528, __extension__ __PRETTY_FUNCTION__));
2529 return MemoryAccessAttr::get(odsParser.getContext(),
2530 ::mlir::spirv::MemoryAccess((*_result_value)));
2531}
2532
2533void MemoryAccessAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2534 ::mlir::Builder odsBuilder(getContext());
2535 odsPrinter << "<";
2536 odsPrinter << stringifyMemoryAccess(getValue());
2537 odsPrinter << ">";
2538}
2539
2540::mlir::spirv::MemoryAccess MemoryAccessAttr::getValue() const {
2541 return getImpl()->value;
2542}
2543
2544} // namespace spirv
2545} // namespace mlir
2546MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::MemoryAccessAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::MemoryAccessAttr>::id = {}; } }
2547namespace mlir {
2548namespace spirv {
2549namespace detail {
2550struct MemoryModelAttrStorage : public ::mlir::AttributeStorage {
2551 using KeyTy = std::tuple<::mlir::spirv::MemoryModel>;
2552 MemoryModelAttrStorage(::mlir::spirv::MemoryModel value) : value(value) {}
2553
2554 KeyTy getAsKey() const {
2555 return KeyTy(value);
2556 }
2557
2558 bool operator==(const KeyTy &tblgenKey) const {
2559 return (value == std::get<0>(tblgenKey));
2560 }
2561
2562 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2563 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2564 }
2565
2566 static MemoryModelAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2567 auto value = std::get<0>(tblgenKey);
2568 return new (allocator.allocate<MemoryModelAttrStorage>()) MemoryModelAttrStorage(value);
2569 }
2570
2571 ::mlir::spirv::MemoryModel value;
2572};
2573} // namespace detail
2574MemoryModelAttr MemoryModelAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::MemoryModel value) {
2575 return Base::get(context, value);
2576}
2577
2578::mlir::Attribute MemoryModelAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2579 ::mlir::Builder odsBuilder(odsParser.getContext());
2580 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2581 (void) odsLoc;
2582 ::mlir::FailureOr<::mlir::spirv::MemoryModel> _result_value;
2583 // Parse literal '<'
2584 if (odsParser.parseLess()) return {};
2585
2586 // Parse variable 'value'
2587 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::MemoryModel> {
2588 auto loc = odsParser.getCurrentLocation();
2589 ::llvm::StringRef enumKeyword;
2590 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2591 return ::mlir::failure();
2592 auto maybeEnum = ::mlir::spirv::symbolizeMemoryModel(enumKeyword);
2593 if (maybeEnum)
2594 return *maybeEnum;
2595 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::MemoryModel" << " to be one of: " << "Simple" << ", " << "GLSL450" << ", " << "OpenCL" << ", " << "Vulkan")};
2596 }();
2597 if (::mlir::failed(_result_value)) {
2598 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_MemoryModelAttr parameter 'value' which is to be a `::mlir::spirv::MemoryModel`");
2599 return {};
2600 }
2601 // Parse literal '>'
2602 if (odsParser.parseGreater()) return {};
2603 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"
, 2603, __extension__ __PRETTY_FUNCTION__));
2604 return MemoryModelAttr::get(odsParser.getContext(),
2605 ::mlir::spirv::MemoryModel((*_result_value)));
2606}
2607
2608void MemoryModelAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2609 ::mlir::Builder odsBuilder(getContext());
2610 odsPrinter << "<";
2611 odsPrinter << stringifyMemoryModel(getValue());
2612 odsPrinter << ">";
2613}
2614
2615::mlir::spirv::MemoryModel MemoryModelAttr::getValue() const {
2616 return getImpl()->value;
2617}
2618
2619} // namespace spirv
2620} // namespace mlir
2621MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::MemoryModelAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::MemoryModelAttr>::id = {}; } }
2622namespace mlir {
2623namespace spirv {
2624namespace detail {
2625struct MemorySemanticsAttrStorage : public ::mlir::AttributeStorage {
2626 using KeyTy = std::tuple<::mlir::spirv::MemorySemantics>;
2627 MemorySemanticsAttrStorage(::mlir::spirv::MemorySemantics value) : value(value) {}
2628
2629 KeyTy getAsKey() const {
2630 return KeyTy(value);
2631 }
2632
2633 bool operator==(const KeyTy &tblgenKey) const {
2634 return (value == std::get<0>(tblgenKey));
2635 }
2636
2637 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2638 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2639 }
2640
2641 static MemorySemanticsAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2642 auto value = std::get<0>(tblgenKey);
2643 return new (allocator.allocate<MemorySemanticsAttrStorage>()) MemorySemanticsAttrStorage(value);
2644 }
2645
2646 ::mlir::spirv::MemorySemantics value;
2647};
2648} // namespace detail
2649MemorySemanticsAttr MemorySemanticsAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::MemorySemantics value) {
2650 return Base::get(context, value);
2651}
2652
2653::mlir::Attribute MemorySemanticsAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2654 ::mlir::Builder odsBuilder(odsParser.getContext());
2655 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2656 (void) odsLoc;
2657 ::mlir::FailureOr<::mlir::spirv::MemorySemantics> _result_value;
2658 // Parse literal '<'
2659 if (odsParser.parseLess()) return {};
2660
2661 // Parse variable 'value'
2662 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::MemorySemantics> {
2663 ::mlir::spirv::MemorySemantics flags = {};
2664 auto loc = odsParser.getCurrentLocation();
2665 ::llvm::StringRef enumKeyword;
2666 do {
2667 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2668 return ::mlir::failure();
2669 auto maybeEnum = ::mlir::spirv::symbolizeMemorySemantics(enumKeyword);
2670 if (!maybeEnum) {
2671 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")};
2672 }
2673 flags = flags | *maybeEnum;
2674 } while(::mlir::succeeded(odsParser.parseOptionalVerticalBar()));
2675 return flags;
2676 }();
2677 if (::mlir::failed(_result_value)) {
2678 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_MemorySemanticsAttr parameter 'value' which is to be a `::mlir::spirv::MemorySemantics`");
2679 return {};
2680 }
2681 // Parse literal '>'
2682 if (odsParser.parseGreater()) return {};
2683 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"
, 2683, __extension__ __PRETTY_FUNCTION__));
2684 return MemorySemanticsAttr::get(odsParser.getContext(),
2685 ::mlir::spirv::MemorySemantics((*_result_value)));
2686}
2687
2688void MemorySemanticsAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2689 ::mlir::Builder odsBuilder(getContext());
2690 odsPrinter << "<";
2691 odsPrinter << stringifyMemorySemantics(getValue());
2692 odsPrinter << ">";
2693}
2694
2695::mlir::spirv::MemorySemantics MemorySemanticsAttr::getValue() const {
2696 return getImpl()->value;
2697}
2698
2699} // namespace spirv
2700} // namespace mlir
2701MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::MemorySemanticsAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::MemorySemanticsAttr>::id = {}; } }
2702namespace mlir {
2703namespace spirv {
2704namespace detail {
2705struct OpcodeAttrStorage : public ::mlir::AttributeStorage {
2706 using KeyTy = std::tuple<::mlir::spirv::Opcode>;
2707 OpcodeAttrStorage(::mlir::spirv::Opcode value) : value(value) {}
2708
2709 KeyTy getAsKey() const {
2710 return KeyTy(value);
2711 }
2712
2713 bool operator==(const KeyTy &tblgenKey) const {
2714 return (value == std::get<0>(tblgenKey));
2715 }
2716
2717 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2718 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2719 }
2720
2721 static OpcodeAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2722 auto value = std::get<0>(tblgenKey);
2723 return new (allocator.allocate<OpcodeAttrStorage>()) OpcodeAttrStorage(value);
2724 }
2725
2726 ::mlir::spirv::Opcode value;
2727};
2728} // namespace detail
2729OpcodeAttr OpcodeAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Opcode value) {
2730 return Base::get(context, value);
2731}
2732
2733::mlir::Attribute OpcodeAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2734 ::mlir::Builder odsBuilder(odsParser.getContext());
2735 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2736 (void) odsLoc;
2737 ::mlir::FailureOr<::mlir::spirv::Opcode> _result_value;
2738 // Parse literal '<'
2739 if (odsParser.parseLess()) return {};
2740
2741 // Parse variable 'value'
2742 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Opcode> {
2743 auto loc = odsParser.getCurrentLocation();
2744 ::llvm::StringRef enumKeyword;
2745 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2746 return ::mlir::failure();
2747 auto maybeEnum = ::mlir::spirv::symbolizeOpcode(enumKeyword);
2748 if (maybeEnum)
2749 return *maybeEnum;
2750 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" << ", " << "OpUMulExtended" << ", " << "OpSMulExtended" << ", " << "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" << ", " << "OpGroupIAdd" << ", " << "OpGroupFAdd" << ", " << "OpGroupFMin" << ", " << "OpGroupUMin" << ", " << "OpGroupSMin" << ", " << "OpGroupFMax" << ", " << "OpGroupUMax" << ", " << "OpGroupSMax" << ", " << "OpNoLine" << ", " << "OpModuleProcessed" << ", " << "OpGroupNonUniformElect" << ", " << "OpGroupNonUniformBroadcast" << ", " << "OpGroupNonUniformBallot" << ", " << "OpGroupNonUniformShuffle" << ", " << "OpGroupNonUniformShuffleXor" << ", " << "OpGroupNonUniformShuffleUp" << ", " << "OpGroupNonUniformShuffleDown" << ", " << "OpGroupNonUniformIAdd" << ", " << "OpGroupNonUniformFAdd" << ", " << "OpGroupNonUniformIMul" << ", " << "OpGroupNonUniformFMul" << ", " << "OpGroupNonUniformSMin" << ", " << "OpGroupNonUniformUMin" << ", " << "OpGroupNonUniformFMin" << ", " << "OpGroupNonUniformSMax" << ", " << "OpGroupNonUniformUMax" << ", " << "OpGroupNonUniformFMax" << ", " << "OpSubgroupBallotKHR" << ", " << "OpSDot" << ", " << "OpUDot" << ", " << "OpSUDot" << ", " << "OpSDotAccSat" << ", " << "OpUDotAccSat" << ", " << "OpSUDotAccSat" << ", " << "OpTypeCooperativeMatrixNV" << ", " << "OpCooperativeMatrixLoadNV" << ", " << "OpCooperativeMatrixStoreNV" << ", " << "OpCooperativeMatrixMulAddNV" << ", " << "OpCooperativeMatrixLengthNV" << ", " << "OpSubgroupBlockReadINTEL" << ", " << "OpSubgroupBlockWriteINTEL" << ", " << "OpAssumeTrueKHR" << ", " << "OpAtomicFAddEXT" << ", " << "OpGroupIMulKHR" << ", " << "OpGroupFMulKHR" << ", " << "OpTypeJointMatrixINTEL" << ", " << "OpJointMatrixLoadINTEL" << ", " << "OpJointMatrixStoreINTEL" << ", " << "OpJointMatrixMadINTEL" << ", " << "OpJointMatrixWorkItemLengthINTEL" << ", " << "OpConvertFToBF16INTEL" << ", " << "OpConvertBF16ToFINTEL")};
2751 }();
2752 if (::mlir::failed(_result_value)) {
2753 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_OpcodeAttr parameter 'value' which is to be a `::mlir::spirv::Opcode`");
2754 return {};
2755 }
2756 // Parse literal '>'
2757 if (odsParser.parseGreater()) return {};
2758 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"
, 2758, __extension__ __PRETTY_FUNCTION__));
2759 return OpcodeAttr::get(odsParser.getContext(),
2760 ::mlir::spirv::Opcode((*_result_value)));
2761}
2762
2763void OpcodeAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2764 ::mlir::Builder odsBuilder(getContext());
2765 odsPrinter << "<";
2766 odsPrinter << stringifyOpcode(getValue());
2767 odsPrinter << ">";
2768}
2769
2770::mlir::spirv::Opcode OpcodeAttr::getValue() const {
2771 return getImpl()->value;
2772}
2773
2774} // namespace spirv
2775} // namespace mlir
2776MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::OpcodeAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::OpcodeAttr>::id = {}; } }
2777namespace mlir {
2778namespace spirv {
2779namespace detail {
2780struct PackedVectorFormatAttrStorage : public ::mlir::AttributeStorage {
2781 using KeyTy = std::tuple<::mlir::spirv::PackedVectorFormat>;
2782 PackedVectorFormatAttrStorage(::mlir::spirv::PackedVectorFormat value) : value(value) {}
2783
2784 KeyTy getAsKey() const {
2785 return KeyTy(value);
2786 }
2787
2788 bool operator==(const KeyTy &tblgenKey) const {
2789 return (value == std::get<0>(tblgenKey));
2790 }
2791
2792 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2793 return ::llvm::hash_combine(std::get<0>(tblgenKey));
2794 }
2795
2796 static PackedVectorFormatAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2797 auto value = std::get<0>(tblgenKey);
2798 return new (allocator.allocate<PackedVectorFormatAttrStorage>()) PackedVectorFormatAttrStorage(value);
2799 }
2800
2801 ::mlir::spirv::PackedVectorFormat value;
2802};
2803} // namespace detail
2804PackedVectorFormatAttr PackedVectorFormatAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::PackedVectorFormat value) {
2805 return Base::get(context, value);
2806}
2807
2808::mlir::Attribute PackedVectorFormatAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2809 ::mlir::Builder odsBuilder(odsParser.getContext());
2810 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2811 (void) odsLoc;
2812 ::mlir::FailureOr<::mlir::spirv::PackedVectorFormat> _result_value;
2813 // Parse literal '<'
2814 if (odsParser.parseLess()) return {};
2815
2816 // Parse variable 'value'
2817 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::PackedVectorFormat> {
2818 auto loc = odsParser.getCurrentLocation();
2819 ::llvm::StringRef enumKeyword;
2820 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
2821 return ::mlir::failure();
2822 auto maybeEnum = ::mlir::spirv::symbolizePackedVectorFormat(enumKeyword);
2823 if (maybeEnum)
2824 return *maybeEnum;
2825 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::PackedVectorFormat" << " to be one of: " << "PackedVectorFormat4x8Bit")};
2826 }();
2827 if (::mlir::failed(_result_value)) {
2828 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_PackedVectorFormatAttr parameter 'value' which is to be a `::mlir::spirv::PackedVectorFormat`");
2829 return {};
2830 }
2831 // Parse literal '>'
2832 if (odsParser.parseGreater()) return {};
2833 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"
, 2833, __extension__ __PRETTY_FUNCTION__));
2834 return PackedVectorFormatAttr::get(odsParser.getContext(),
2835 ::mlir::spirv::PackedVectorFormat((*_result_value)));
2836}
2837
2838void PackedVectorFormatAttr::print(::mlir::AsmPrinter &odsPrinter) const {
2839 ::mlir::Builder odsBuilder(getContext());
2840 odsPrinter << "<";
2841 odsPrinter << stringifyPackedVectorFormat(getValue());
2842 odsPrinter << ">";
2843}
2844
2845::mlir::spirv::PackedVectorFormat PackedVectorFormatAttr::getValue() const {
2846 return getImpl()->value;
2847}
2848
2849} // namespace spirv
2850} // namespace mlir
2851MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::PackedVectorFormatAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::PackedVectorFormatAttr>::id = {}; } }
2852namespace mlir {
2853namespace spirv {
2854namespace detail {
2855struct ResourceLimitsAttrStorage : public ::mlir::AttributeStorage {
2856 using KeyTy = std::tuple<int, int, ArrayAttr, int, std::optional<int>, std::optional<int>, ArrayAttr>;
2857 ResourceLimitsAttrStorage(int max_compute_shared_memory_size, int max_compute_workgroup_invocations, ArrayAttr max_compute_workgroup_size, int subgroup_size, std::optional<int> min_subgroup_size, std::optional<int> max_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), min_subgroup_size(min_subgroup_size), max_subgroup_size(max_subgroup_size), cooperative_matrix_properties_nv(cooperative_matrix_properties_nv) {}
2858
2859 KeyTy getAsKey() const {
2860 return KeyTy(max_compute_shared_memory_size, max_compute_workgroup_invocations, max_compute_workgroup_size, subgroup_size, min_subgroup_size, max_subgroup_size, cooperative_matrix_properties_nv);
2861 }
2862
2863 bool operator==(const KeyTy &tblgenKey) const {
2864 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)) && (min_subgroup_size == std::get<4>(tblgenKey)) && (max_subgroup_size == std::get<5>(tblgenKey)) && (cooperative_matrix_properties_nv == std::get<6>(tblgenKey));
2865 }
2866
2867 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
2868 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));
2869 }
2870
2871 static ResourceLimitsAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
2872 auto max_compute_shared_memory_size = std::get<0>(tblgenKey);
2873 auto max_compute_workgroup_invocations = std::get<1>(tblgenKey);
2874 auto max_compute_workgroup_size = std::get<2>(tblgenKey);
2875 auto subgroup_size = std::get<3>(tblgenKey);
2876 auto min_subgroup_size = std::get<4>(tblgenKey);
2877 auto max_subgroup_size = std::get<5>(tblgenKey);
2878 auto cooperative_matrix_properties_nv = std::get<6>(tblgenKey);
2879 return new (allocator.allocate<ResourceLimitsAttrStorage>()) ResourceLimitsAttrStorage(max_compute_shared_memory_size, max_compute_workgroup_invocations, max_compute_workgroup_size, subgroup_size, min_subgroup_size, max_subgroup_size, cooperative_matrix_properties_nv);
2880 }
2881
2882 int max_compute_shared_memory_size;
2883 int max_compute_workgroup_invocations;
2884 ArrayAttr max_compute_workgroup_size;
2885 int subgroup_size;
2886 std::optional<int> min_subgroup_size;
2887 std::optional<int> max_subgroup_size;
2888 ArrayAttr cooperative_matrix_properties_nv;
2889};
2890} // namespace detail
2891ResourceLimitsAttr ResourceLimitsAttr::get(::mlir::MLIRContext *context, int max_compute_shared_memory_size, int max_compute_workgroup_invocations, ArrayAttr max_compute_workgroup_size, int subgroup_size, std::optional<int> min_subgroup_size, std::optional<int> max_subgroup_size, ArrayAttr cooperative_matrix_properties_nv) {
2892 return Base::get(context, max_compute_shared_memory_size, max_compute_workgroup_invocations, max_compute_workgroup_size, subgroup_size, min_subgroup_size, max_subgroup_size, cooperative_matrix_properties_nv);
2893}
2894
2895::mlir::Attribute ResourceLimitsAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
2896 ::mlir::Builder odsBuilder(odsParser.getContext());
2897 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
2898 (void) odsLoc;
2899 ::mlir::FailureOr<int> _result_max_compute_shared_memory_size;
2900 ::mlir::FailureOr<int> _result_max_compute_workgroup_invocations;
2901 ::mlir::FailureOr<ArrayAttr> _result_max_compute_workgroup_size;
2902 ::mlir::FailureOr<int> _result_subgroup_size;
2903 ::mlir::FailureOr<std::optional<int>> _result_min_subgroup_size;
2904 ::mlir::FailureOr<std::optional<int>> _result_max_subgroup_size;
2905 ::mlir::FailureOr<ArrayAttr> _result_cooperative_matrix_properties_nv;
2906 // Parse literal '<'
2907 if (odsParser.parseLess()) return {};
2908 // Parse parameter struct
2909 bool _seen_max_compute_shared_memory_size = false;
2910 bool _seen_max_compute_workgroup_invocations = false;
2911 bool _seen_max_compute_workgroup_size = false;
2912 bool _seen_subgroup_size = false;
2913 bool _seen_min_subgroup_size = false;
2914 bool _seen_max_subgroup_size = false;
2915 bool _seen_cooperative_matrix_properties_nv = false;
2916 {
2917 const auto _loop_body = [&](::llvm::StringRef _paramKey) -> bool {
2918 // Parse literal '='
2919 if (odsParser.parseEqual()) return {};
2920 if (!_seen_max_compute_shared_memory_size && _paramKey == "max_compute_shared_memory_size") {
2921 _seen_max_compute_shared_memory_size = true;
2922
2923 // Parse variable 'max_compute_shared_memory_size'
2924 _result_max_compute_shared_memory_size = ::mlir::FieldParser<int>::parse(odsParser);
2925 if (::mlir::failed(_result_max_compute_shared_memory_size)) {
2926 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ResourceLimitsAttr parameter 'max_compute_shared_memory_size' which is to be a `int`");
2927 return {};
2928 }
2929 } else if (!_seen_max_compute_workgroup_invocations && _paramKey == "max_compute_workgroup_invocations") {
2930 _seen_max_compute_workgroup_invocations = true;
2931
2932 // Parse variable 'max_compute_workgroup_invocations'
2933 _result_max_compute_workgroup_invocations = ::mlir::FieldParser<int>::parse(odsParser);
2934 if (::mlir::failed(_result_max_compute_workgroup_invocations)) {
2935 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ResourceLimitsAttr parameter 'max_compute_workgroup_invocations' which is to be a `int`");
2936 return {};
2937 }
2938 } else if (!_seen_max_compute_workgroup_size && _paramKey == "max_compute_workgroup_size") {
2939 _seen_max_compute_workgroup_size = true;
2940
2941 // Parse variable 'max_compute_workgroup_size'
2942 _result_max_compute_workgroup_size = ::mlir::FieldParser<ArrayAttr>::parse(odsParser);
2943 if (::mlir::failed(_result_max_compute_workgroup_size)) {
2944 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ResourceLimitsAttr parameter 'max_compute_workgroup_size' which is to be a `ArrayAttr`");
2945 return {};
2946 }
2947 } else if (!_seen_subgroup_size && _paramKey == "subgroup_size") {
2948 _seen_subgroup_size = true;
2949
2950 // Parse variable 'subgroup_size'
2951 _result_subgroup_size = ::mlir::FieldParser<int>::parse(odsParser);
2952 if (::mlir::failed(_result_subgroup_size)) {
2953 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ResourceLimitsAttr parameter 'subgroup_size' which is to be a `int`");
2954 return {};
2955 }
2956 } else if (!_seen_min_subgroup_size && _paramKey == "min_subgroup_size") {
2957 _seen_min_subgroup_size = true;
2958
2959 // Parse variable 'min_subgroup_size'
2960 _result_min_subgroup_size = ::mlir::FieldParser<std::optional<int>>::parse(odsParser);
2961 if (::mlir::failed(_result_min_subgroup_size)) {
2962 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ResourceLimitsAttr parameter 'min_subgroup_size' which is to be a `std::optional<int>`");
2963 return {};
2964 }
2965 } else if (!_seen_max_subgroup_size && _paramKey == "max_subgroup_size") {
2966 _seen_max_subgroup_size = true;
2967
2968 // Parse variable 'max_subgroup_size'
2969 _result_max_subgroup_size = ::mlir::FieldParser<std::optional<int>>::parse(odsParser);
2970 if (::mlir::failed(_result_max_subgroup_size)) {
2971 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ResourceLimitsAttr parameter 'max_subgroup_size' which is to be a `std::optional<int>`");
2972 return {};
2973 }
2974 } else if (!_seen_cooperative_matrix_properties_nv && _paramKey == "cooperative_matrix_properties_nv") {
2975 _seen_cooperative_matrix_properties_nv = true;
2976
2977 // Parse variable 'cooperative_matrix_properties_nv'
2978 _result_cooperative_matrix_properties_nv = ::mlir::FieldParser<ArrayAttr>::parse(odsParser);
2979 if (::mlir::failed(_result_cooperative_matrix_properties_nv)) {
2980 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ResourceLimitsAttr parameter 'cooperative_matrix_properties_nv' which is to be a `ArrayAttr`");
2981 return {};
2982 }
2983 } else {
2984 odsParser.emitError(odsParser.getCurrentLocation(), "duplicate or unknown struct parameter name: ") << _paramKey;
2985 return {};
2986 }
2987 return true;
2988 };
2989 ::llvm::StringRef _paramKey;
2990 if (!odsParser.parseOptionalKeyword(&_paramKey)) {
2991 if (!_loop_body(_paramKey)) return {};
2992 while (!odsParser.parseOptionalComma()) {
2993 ::llvm::StringRef _paramKey;
2994 if (odsParser.parseKeyword(&_paramKey)) {
2995 odsParser.emitError(odsParser.getCurrentLocation(),
2996 "expected a parameter name in struct");
2997 return {};
2998 }
2999 if (!_loop_body(_paramKey)) return {};
3000 }
3001 }
3002 }
3003 // Parse literal '>'
3004 if (odsParser.parseGreater()) return {};
3005 return ResourceLimitsAttr::get(odsParser.getContext(),
3006 int((_result_max_compute_shared_memory_size.value_or(16384))),
3007 int((_result_max_compute_workgroup_invocations.value_or(128))),
3008 ArrayAttr((_result_max_compute_workgroup_size.value_or(odsBuilder.getI32ArrayAttr({128, 128, 64})))),
3009 int((_result_subgroup_size.value_or(32))),
3010 std::optional<int>((_result_min_subgroup_size.value_or(std::optional<int>()))),
3011 std::optional<int>((_result_max_subgroup_size.value_or(std::optional<int>()))),
3012 ArrayAttr((_result_cooperative_matrix_properties_nv.value_or(nullptr))));
3013}
3014
3015void ResourceLimitsAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3016 ::mlir::Builder odsBuilder(getContext());
3017 odsPrinter << "<";
3018 {
3019 bool _firstPrinted = true;
3020 if (!(getMaxComputeSharedMemorySize() == 16384)) {
3021 if (!_firstPrinted) odsPrinter << ", ";
3022 _firstPrinted = false;
3023 odsPrinter << "max_compute_shared_memory_size = ";
3024 if (!(getMaxComputeSharedMemorySize() == 16384)) {
3025 odsPrinter.printStrippedAttrOrType(getMaxComputeSharedMemorySize());
3026 }
3027 }
3028 if (!(getMaxComputeWorkgroupInvocations() == 128)) {
3029 if (!_firstPrinted) odsPrinter << ", ";
3030 _firstPrinted = false;
3031 odsPrinter << "max_compute_workgroup_invocations = ";
3032 if (!(getMaxComputeWorkgroupInvocations() == 128)) {
3033 odsPrinter.printStrippedAttrOrType(getMaxComputeWorkgroupInvocations());
3034 }
3035 }
3036 if (!(getMaxComputeWorkgroupSize() == odsBuilder.getI32ArrayAttr({128, 128, 64}))) {
3037 if (!_firstPrinted) odsPrinter << ", ";
3038 _firstPrinted = false;
3039 odsPrinter << "max_compute_workgroup_size = ";
3040 if (!(getMaxComputeWorkgroupSize() == odsBuilder.getI32ArrayAttr({128, 128, 64}))) {
3041 odsPrinter.printStrippedAttrOrType(getMaxComputeWorkgroupSize());
3042 }
3043 }
3044 if (!(getSubgroupSize() == 32)) {
3045 if (!_firstPrinted) odsPrinter << ", ";
3046 _firstPrinted = false;
3047 odsPrinter << "subgroup_size = ";
3048 if (!(getSubgroupSize() == 32)) {
3049 odsPrinter.printStrippedAttrOrType(getSubgroupSize());
3050 }
3051 }
3052 if (!(getMinSubgroupSize() == std::optional<int>())) {
3053 if (!_firstPrinted) odsPrinter << ", ";
3054 _firstPrinted = false;
3055 odsPrinter << "min_subgroup_size = ";
3056 if (!(getMinSubgroupSize() == std::optional<int>())) {
3057 odsPrinter.printStrippedAttrOrType(getMinSubgroupSize());
3058 }
3059 }
3060 if (!(getMaxSubgroupSize() == std::optional<int>())) {
3061 if (!_firstPrinted) odsPrinter << ", ";
3062 _firstPrinted = false;
3063 odsPrinter << "max_subgroup_size = ";
3064 if (!(getMaxSubgroupSize() == std::optional<int>())) {
3065 odsPrinter.printStrippedAttrOrType(getMaxSubgroupSize());
3066 }
3067 }
3068 if (!(getCooperativeMatrixPropertiesNv() == nullptr)) {
3069 if (!_firstPrinted) odsPrinter << ", ";
3070 _firstPrinted = false;
3071 odsPrinter << "cooperative_matrix_properties_nv = ";
3072 if (!(getCooperativeMatrixPropertiesNv() == nullptr)) {
3073 odsPrinter.printStrippedAttrOrType(getCooperativeMatrixPropertiesNv());
3074 }
3075 }
3076 }
3077 odsPrinter << ">";
3078}
3079
3080int ResourceLimitsAttr::getMaxComputeSharedMemorySize() const {
3081 return getImpl()->max_compute_shared_memory_size;
3082}
3083
3084int ResourceLimitsAttr::getMaxComputeWorkgroupInvocations() const {
3085 return getImpl()->max_compute_workgroup_invocations;
3086}
3087
3088ArrayAttr ResourceLimitsAttr::getMaxComputeWorkgroupSize() const {
3089 return getImpl()->max_compute_workgroup_size;
3090}
3091
3092int ResourceLimitsAttr::getSubgroupSize() const {
3093 return getImpl()->subgroup_size;
3094}
3095
3096std::optional<int> ResourceLimitsAttr::getMinSubgroupSize() const {
3097 return getImpl()->min_subgroup_size;
3098}
3099
3100std::optional<int> ResourceLimitsAttr::getMaxSubgroupSize() const {
3101 return getImpl()->max_subgroup_size;
3102}
3103
3104ArrayAttr ResourceLimitsAttr::getCooperativeMatrixPropertiesNv() const {
3105 return getImpl()->cooperative_matrix_properties_nv;
3106}
3107
3108} // namespace spirv
3109} // namespace mlir
3110MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ResourceLimitsAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ResourceLimitsAttr>::id = {}; } }
3111namespace mlir {
3112namespace spirv {
3113namespace detail {
3114struct ImageSamplerUseInfoAttrStorage : public ::mlir::AttributeStorage {
3115 using KeyTy = std::tuple<::mlir::spirv::ImageSamplerUseInfo>;
3116 ImageSamplerUseInfoAttrStorage(::mlir::spirv::ImageSamplerUseInfo value) : value(value) {}
3117
3118 KeyTy getAsKey() const {
3119 return KeyTy(value);
3120 }
3121
3122 bool operator==(const KeyTy &tblgenKey) const {
3123 return (value == std::get<0>(tblgenKey));
3124 }
3125
3126 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
3127 return ::llvm::hash_combine(std::get<0>(tblgenKey));
3128 }
3129
3130 static ImageSamplerUseInfoAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
3131 auto value = std::get<0>(tblgenKey);
3132 return new (allocator.allocate<ImageSamplerUseInfoAttrStorage>()) ImageSamplerUseInfoAttrStorage(value);
3133 }
3134
3135 ::mlir::spirv::ImageSamplerUseInfo value;
3136};
3137} // namespace detail
3138ImageSamplerUseInfoAttr ImageSamplerUseInfoAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ImageSamplerUseInfo value) {
3139 return Base::get(context, value);
3140}
3141
3142::mlir::Attribute ImageSamplerUseInfoAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
3143 ::mlir::Builder odsBuilder(odsParser.getContext());
3144 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
3145 (void) odsLoc;
3146 ::mlir::FailureOr<::mlir::spirv::ImageSamplerUseInfo> _result_value;
3147 // Parse literal '<'
3148 if (odsParser.parseLess()) return {};
3149
3150 // Parse variable 'value'
3151 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ImageSamplerUseInfo> {
3152 auto loc = odsParser.getCurrentLocation();
3153 ::llvm::StringRef enumKeyword;
3154 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
3155 return ::mlir::failure();
3156 auto maybeEnum = ::mlir::spirv::symbolizeImageSamplerUseInfo(enumKeyword);
3157 if (maybeEnum)
3158 return *maybeEnum;
3159 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ImageSamplerUseInfo" << " to be one of: " << "SamplerUnknown" << ", " << "NeedSampler" << ", " << "NoSampler")};
3160 }();
3161 if (::mlir::failed(_result_value)) {
3162 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_SamplerUseAttr parameter 'value' which is to be a `::mlir::spirv::ImageSamplerUseInfo`");
3163 return {};
3164 }
3165 // Parse literal '>'
3166 if (odsParser.parseGreater()) return {};
3167 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"
, 3167, __extension__ __PRETTY_FUNCTION__));
3168 return ImageSamplerUseInfoAttr::get(odsParser.getContext(),
3169 ::mlir::spirv::ImageSamplerUseInfo((*_result_value)));
3170}
3171
3172void ImageSamplerUseInfoAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3173 ::mlir::Builder odsBuilder(getContext());
3174 odsPrinter << "<";
3175 odsPrinter << stringifyImageSamplerUseInfo(getValue());
3176 odsPrinter << ">";
3177}
3178
3179::mlir::spirv::ImageSamplerUseInfo ImageSamplerUseInfoAttr::getValue() const {
3180 return getImpl()->value;
3181}
3182
3183} // namespace spirv
3184} // namespace mlir
3185MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ImageSamplerUseInfoAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ImageSamplerUseInfoAttr>::id = {}; } }
3186namespace mlir {
3187namespace spirv {
3188namespace detail {
3189struct ImageSamplingInfoAttrStorage : public ::mlir::AttributeStorage {
3190 using KeyTy = std::tuple<::mlir::spirv::ImageSamplingInfo>;
3191 ImageSamplingInfoAttrStorage(::mlir::spirv::ImageSamplingInfo value) : value(value) {}
3192
3193 KeyTy getAsKey() const {
3194 return KeyTy(value);
3195 }
3196
3197 bool operator==(const KeyTy &tblgenKey) const {
3198 return (value == std::get<0>(tblgenKey));
3199 }
3200
3201 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
3202 return ::llvm::hash_combine(std::get<0>(tblgenKey));
3203 }
3204
3205 static ImageSamplingInfoAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
3206 auto value = std::get<0>(tblgenKey);
3207 return new (allocator.allocate<ImageSamplingInfoAttrStorage>()) ImageSamplingInfoAttrStorage(value);
3208 }
3209
3210 ::mlir::spirv::ImageSamplingInfo value;
3211};
3212} // namespace detail
3213ImageSamplingInfoAttr ImageSamplingInfoAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::ImageSamplingInfo value) {
3214 return Base::get(context, value);
3215}
3216
3217::mlir::Attribute ImageSamplingInfoAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
3218 ::mlir::Builder odsBuilder(odsParser.getContext());
3219 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
3220 (void) odsLoc;
3221 ::mlir::FailureOr<::mlir::spirv::ImageSamplingInfo> _result_value;
3222 // Parse literal '<'
3223 if (odsParser.parseLess()) return {};
3224
3225 // Parse variable 'value'
3226 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::ImageSamplingInfo> {
3227 auto loc = odsParser.getCurrentLocation();
3228 ::llvm::StringRef enumKeyword;
3229 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
3230 return ::mlir::failure();
3231 auto maybeEnum = ::mlir::spirv::symbolizeImageSamplingInfo(enumKeyword);
3232 if (maybeEnum)
3233 return *maybeEnum;
3234 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::ImageSamplingInfo" << " to be one of: " << "SingleSampled" << ", " << "MultiSampled")};
3235 }();
3236 if (::mlir::failed(_result_value)) {
3237 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_SamplingAttr parameter 'value' which is to be a `::mlir::spirv::ImageSamplingInfo`");
3238 return {};
3239 }
3240 // Parse literal '>'
3241 if (odsParser.parseGreater()) return {};
3242 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"
, 3242, __extension__ __PRETTY_FUNCTION__));
3243 return ImageSamplingInfoAttr::get(odsParser.getContext(),
3244 ::mlir::spirv::ImageSamplingInfo((*_result_value)));
3245}
3246
3247void ImageSamplingInfoAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3248 ::mlir::Builder odsBuilder(getContext());
3249 odsPrinter << "<";
3250 odsPrinter << stringifyImageSamplingInfo(getValue());
3251 odsPrinter << ">";
3252}
3253
3254::mlir::spirv::ImageSamplingInfo ImageSamplingInfoAttr::getValue() const {
3255 return getImpl()->value;
3256}
3257
3258} // namespace spirv
3259} // namespace mlir
3260MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ImageSamplingInfoAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ImageSamplingInfoAttr>::id = {}; } }
3261namespace mlir {
3262namespace spirv {
3263namespace detail {
3264struct ScopeAttrStorage : public ::mlir::AttributeStorage {
3265 using KeyTy = std::tuple<::mlir::spirv::Scope>;
3266 ScopeAttrStorage(::mlir::spirv::Scope value) : value(value) {}
3267
3268 KeyTy getAsKey() const {
3269 return KeyTy(value);
3270 }
3271
3272 bool operator==(const KeyTy &tblgenKey) const {
3273 return (value == std::get<0>(tblgenKey));
3274 }
3275
3276 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
3277 return ::llvm::hash_combine(std::get<0>(tblgenKey));
3278 }
3279
3280 static ScopeAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
3281 auto value = std::get<0>(tblgenKey);
3282 return new (allocator.allocate<ScopeAttrStorage>()) ScopeAttrStorage(value);
3283 }
3284
3285 ::mlir::spirv::Scope value;
3286};
3287} // namespace detail
3288ScopeAttr ScopeAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Scope value) {
3289 return Base::get(context, value);
3290}
3291
3292::mlir::Attribute ScopeAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
3293 ::mlir::Builder odsBuilder(odsParser.getContext());
3294 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
3295 (void) odsLoc;
3296 ::mlir::FailureOr<::mlir::spirv::Scope> _result_value;
3297 // Parse literal '<'
3298 if (odsParser.parseLess()) return {};
3299
3300 // Parse variable 'value'
3301 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Scope> {
3302 auto loc = odsParser.getCurrentLocation();
3303 ::llvm::StringRef enumKeyword;
3304 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
3305 return ::mlir::failure();
3306 auto maybeEnum = ::mlir::spirv::symbolizeScope(enumKeyword);
3307 if (maybeEnum)
3308 return *maybeEnum;
3309 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::Scope" << " to be one of: " << "CrossDevice" << ", " << "Device" << ", " << "Workgroup" << ", " << "Subgroup" << ", " << "Invocation" << ", " << "QueueFamily" << ", " << "ShaderCallKHR")};
3310 }();
3311 if (::mlir::failed(_result_value)) {
3312 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_ScopeAttr parameter 'value' which is to be a `::mlir::spirv::Scope`");
3313 return {};
3314 }
3315 // Parse literal '>'
3316 if (odsParser.parseGreater()) return {};
3317 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"
, 3317, __extension__ __PRETTY_FUNCTION__));
3318 return ScopeAttr::get(odsParser.getContext(),
3319 ::mlir::spirv::Scope((*_result_value)));
3320}
3321
3322void ScopeAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3323 ::mlir::Builder odsBuilder(getContext());
3324 odsPrinter << "<";
3325 odsPrinter << stringifyScope(getValue());
3326 odsPrinter << ">";
3327}
3328
3329::mlir::spirv::Scope ScopeAttr::getValue() const {
3330 return getImpl()->value;
3331}
3332
3333} // namespace spirv
3334} // namespace mlir
3335MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::ScopeAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::ScopeAttr>::id = {}; } }
3336namespace mlir {
3337namespace spirv {
3338namespace detail {
3339struct SelectionControlAttrStorage : public ::mlir::AttributeStorage {
3340 using KeyTy = std::tuple<::mlir::spirv::SelectionControl>;
3341 SelectionControlAttrStorage(::mlir::spirv::SelectionControl value) : value(value) {}
3342
3343 KeyTy getAsKey() const {
3344 return KeyTy(value);
3345 }
3346
3347 bool operator==(const KeyTy &tblgenKey) const {
3348 return (value == std::get<0>(tblgenKey));
3349 }
3350
3351 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
3352 return ::llvm::hash_combine(std::get<0>(tblgenKey));
3353 }
3354
3355 static SelectionControlAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
3356 auto value = std::get<0>(tblgenKey);
3357 return new (allocator.allocate<SelectionControlAttrStorage>()) SelectionControlAttrStorage(value);
3358 }
3359
3360 ::mlir::spirv::SelectionControl value;
3361};
3362} // namespace detail
3363SelectionControlAttr SelectionControlAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::SelectionControl value) {
3364 return Base::get(context, value);
3365}
3366
3367::mlir::Attribute SelectionControlAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
3368 ::mlir::Builder odsBuilder(odsParser.getContext());
3369 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
3370 (void) odsLoc;
3371 ::mlir::FailureOr<::mlir::spirv::SelectionControl> _result_value;
3372 // Parse literal '<'
3373 if (odsParser.parseLess()) return {};
3374
3375 // Parse variable 'value'
3376 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::SelectionControl> {
3377 ::mlir::spirv::SelectionControl flags = {};
3378 auto loc = odsParser.getCurrentLocation();
3379 ::llvm::StringRef enumKeyword;
3380 do {
3381 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
3382 return ::mlir::failure();
3383 auto maybeEnum = ::mlir::spirv::symbolizeSelectionControl(enumKeyword);
3384 if (!maybeEnum) {
3385 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::SelectionControl" << " to be one of: " << "None" << ", " << "Flatten" << ", " << "DontFlatten")};
3386 }
3387 flags = flags | *maybeEnum;
3388 } while(::mlir::succeeded(odsParser.parseOptionalVerticalBar()));
3389 return flags;
3390 }();
3391 if (::mlir::failed(_result_value)) {
3392 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_SelectionControlAttr parameter 'value' which is to be a `::mlir::spirv::SelectionControl`");
3393 return {};
3394 }
3395 // Parse literal '>'
3396 if (odsParser.parseGreater()) return {};
3397 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"
, 3397, __extension__ __PRETTY_FUNCTION__));
3398 return SelectionControlAttr::get(odsParser.getContext(),
3399 ::mlir::spirv::SelectionControl((*_result_value)));
3400}
3401
3402void SelectionControlAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3403 ::mlir::Builder odsBuilder(getContext());
3404 odsPrinter << "<";
3405 odsPrinter << stringifySelectionControl(getValue());
3406 odsPrinter << ">";
3407}
3408
3409::mlir::spirv::SelectionControl SelectionControlAttr::getValue() const {
3410 return getImpl()->value;
3411}
3412
3413} // namespace spirv
3414} // namespace mlir
3415MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::SelectionControlAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::SelectionControlAttr>::id = {}; } }
3416namespace mlir {
3417namespace spirv {
3418namespace detail {
3419struct StorageClassAttrStorage : public ::mlir::AttributeStorage {
3420 using KeyTy = std::tuple<::mlir::spirv::StorageClass>;
3421 StorageClassAttrStorage(::mlir::spirv::StorageClass value) : value(value) {}
3422
3423 KeyTy getAsKey() const {
3424 return KeyTy(value);
3425 }
3426
3427 bool operator==(const KeyTy &tblgenKey) const {
3428 return (value == std::get<0>(tblgenKey));
3429 }
3430
3431 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
3432 return ::llvm::hash_combine(std::get<0>(tblgenKey));
3433 }
3434
3435 static StorageClassAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
3436 auto value = std::get<0>(tblgenKey);
3437 return new (allocator.allocate<StorageClassAttrStorage>()) StorageClassAttrStorage(value);
3438 }
3439
3440 ::mlir::spirv::StorageClass value;
3441};
3442} // namespace detail
3443StorageClassAttr StorageClassAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::StorageClass value) {
3444 return Base::get(context, value);
3445}
3446
3447::mlir::Attribute StorageClassAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
3448 ::mlir::Builder odsBuilder(odsParser.getContext());
3449 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
3450 (void) odsLoc;
3451 ::mlir::FailureOr<::mlir::spirv::StorageClass> _result_value;
3452 // Parse literal '<'
3453 if (odsParser.parseLess()) return {};
3454
3455 // Parse variable 'value'
3456 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::StorageClass> {
3457 auto loc = odsParser.getCurrentLocation();
3458 ::llvm::StringRef enumKeyword;
3459 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
3460 return ::mlir::failure();
3461 auto maybeEnum = ::mlir::spirv::symbolizeStorageClass(enumKeyword);
3462 if (maybeEnum)
3463 return *maybeEnum;
3464 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")};
3465 }();
3466 if (::mlir::failed(_result_value)) {
3467 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_StorageClassAttr parameter 'value' which is to be a `::mlir::spirv::StorageClass`");
3468 return {};
3469 }
3470 // Parse literal '>'
3471 if (odsParser.parseGreater()) return {};
3472 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"
, 3472, __extension__ __PRETTY_FUNCTION__));
3473 return StorageClassAttr::get(odsParser.getContext(),
3474 ::mlir::spirv::StorageClass((*_result_value)));
3475}
3476
3477void StorageClassAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3478 ::mlir::Builder odsBuilder(getContext());
3479 odsPrinter << "<";
3480 odsPrinter << stringifyStorageClass(getValue());
3481 odsPrinter << ">";
3482}
3483
3484::mlir::spirv::StorageClass StorageClassAttr::getValue() const {
3485 return getImpl()->value;
3486}
3487
3488} // namespace spirv
3489} // namespace mlir
3490MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::StorageClassAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::StorageClassAttr>::id = {}; } }
3491namespace mlir {
3492namespace spirv {
3493namespace detail {
3494struct VendorAttrStorage : public ::mlir::AttributeStorage {
3495 using KeyTy = std::tuple<::mlir::spirv::Vendor>;
3496 VendorAttrStorage(::mlir::spirv::Vendor value) : value(value) {}
3497
3498 KeyTy getAsKey() const {
3499 return KeyTy(value);
3500 }
3501
3502 bool operator==(const KeyTy &tblgenKey) const {
3503 return (value == std::get<0>(tblgenKey));
3504 }
3505
3506 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
3507 return ::llvm::hash_combine(std::get<0>(tblgenKey));
3508 }
3509
3510 static VendorAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
3511 auto value = std::get<0>(tblgenKey);
3512 return new (allocator.allocate<VendorAttrStorage>()) VendorAttrStorage(value);
3513 }
3514
3515 ::mlir::spirv::Vendor value;
3516};
3517} // namespace detail
3518VendorAttr VendorAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Vendor value) {
3519 return Base::get(context, value);
3520}
3521
3522::mlir::Attribute VendorAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
3523 ::mlir::Builder odsBuilder(odsParser.getContext());
3524 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
3525 (void) odsLoc;
3526 ::mlir::FailureOr<::mlir::spirv::Vendor> _result_value;
3527 // Parse literal '<'
3528 if (odsParser.parseLess()) return {};
3529
3530 // Parse variable 'value'
3531 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Vendor> {
3532 auto loc = odsParser.getCurrentLocation();
3533 ::llvm::StringRef enumKeyword;
3534 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
3535 return ::mlir::failure();
3536 auto maybeEnum = ::mlir::spirv::symbolizeVendor(enumKeyword);
3537 if (maybeEnum)
3538 return *maybeEnum;
3539 return {(::mlir::LogicalResult)(odsParser.emitError(loc) << "expected " << "::mlir::spirv::Vendor" << " to be one of: " << "AMD" << ", " << "Apple" << ", " << "ARM" << ", " << "Imagination" << ", " << "Intel" << ", " << "NVIDIA" << ", " << "Qualcomm" << ", " << "SwiftShader" << ", " << "Unknown")};
3540 }();
3541 if (::mlir::failed(_result_value)) {
3542 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_VendorAttr parameter 'value' which is to be a `::mlir::spirv::Vendor`");
3543 return {};
3544 }
3545 // Parse literal '>'
3546 if (odsParser.parseGreater()) return {};
3547 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"
, 3547, __extension__ __PRETTY_FUNCTION__));
3548 return VendorAttr::get(odsParser.getContext(),
3549 ::mlir::spirv::Vendor((*_result_value)));
3550}
3551
3552void VendorAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3553 ::mlir::Builder odsBuilder(getContext());
3554 odsPrinter << "<";
3555 odsPrinter << stringifyVendor(getValue());
3556 odsPrinter << ">";
3557}
3558
3559::mlir::spirv::Vendor VendorAttr::getValue() const {
3560 return getImpl()->value;
3561}
3562
3563} // namespace spirv
3564} // namespace mlir
3565MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::VendorAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::VendorAttr>::id = {}; } }
3566namespace mlir {
3567namespace spirv {
3568namespace detail {
3569struct VersionAttrStorage : public ::mlir::AttributeStorage {
3570 using KeyTy = std::tuple<::mlir::spirv::Version>;
3571 VersionAttrStorage(::mlir::spirv::Version value) : value(value) {}
3572
3573 KeyTy getAsKey() const {
3574 return KeyTy(value);
3575 }
3576
3577 bool operator==(const KeyTy &tblgenKey) const {
3578 return (value == std::get<0>(tblgenKey));
3579 }
3580
3581 static ::llvm::hash_code hashKey(const KeyTy &tblgenKey) {
3582 return ::llvm::hash_combine(std::get<0>(tblgenKey));
3583 }
3584
3585 static VersionAttrStorage *construct(::mlir::AttributeStorageAllocator &allocator, const KeyTy &tblgenKey) {
3586 auto value = std::get<0>(tblgenKey);
3587 return new (allocator.allocate<VersionAttrStorage>()) VersionAttrStorage(value);
3588 }
3589
3590 ::mlir::spirv::Version value;
3591};
3592} // namespace detail
3593VersionAttr VersionAttr::get(::mlir::MLIRContext *context, ::mlir::spirv::Version value) {
3594 return Base::get(context, value);
3595}
3596
3597::mlir::Attribute VersionAttr::parse(::mlir::AsmParser &odsParser, ::mlir::Type odsType) {
3598 ::mlir::Builder odsBuilder(odsParser.getContext());
3599 ::llvm::SMLoc odsLoc = odsParser.getCurrentLocation();
3600 (void) odsLoc;
3601 ::mlir::FailureOr<::mlir::spirv::Version> _result_value;
3602 // Parse literal '<'
3603 if (odsParser.parseLess()) return {};
3604
3605 // Parse variable 'value'
3606 _result_value = [&]() -> ::mlir::FailureOr<::mlir::spirv::Version> {
3607 auto loc = odsParser.getCurrentLocation();
3608 ::llvm::StringRef enumKeyword;
3609 if (::mlir::failed(odsParser.parseKeyword(&enumKeyword)))
3610 return ::mlir::failure();
3611 auto maybeEnum = ::mlir::spirv::symbolizeVersion(enumKeyword);
3612 if (maybeEnum)
3613 return *maybeEnum;
3614 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")};
3615 }();
3616 if (::mlir::failed(_result_value)) {
3617 odsParser.emitError(odsParser.getCurrentLocation(), "failed to parse SPIRV_VersionAttr parameter 'value' which is to be a `::mlir::spirv::Version`");
3618 return {};
3619 }
3620 // Parse literal '>'
3621 if (odsParser.parseGreater()) return {};
3622 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"
, 3622, __extension__ __PRETTY_FUNCTION__));
3623 return VersionAttr::get(odsParser.getContext(),
3624 ::mlir::spirv::Version((*_result_value)));
3625}
3626
3627void VersionAttr::print(::mlir::AsmPrinter &odsPrinter) const {
3628 ::mlir::Builder odsBuilder(getContext());
3629 odsPrinter << "<";
3630 odsPrinter << stringifyVersion(getValue());
3631 odsPrinter << ">";
3632}
3633
3634::mlir::spirv::Version VersionAttr::getValue() const {
3635 return getImpl()->value;
3636}
3637
3638} // namespace spirv
3639} // namespace mlir
3640MLIR_DEFINE_EXPLICIT_TYPE_ID(::mlir::spirv::VersionAttr)namespace mlir { namespace detail { SelfOwningTypeID TypeIDResolver
< ::mlir::spirv::VersionAttr>::id = {}; } }
3641
3642#endif // GET_ATTRDEF_CLASSES
3643