Bug Summary

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