clang 19.0.0git
RISCVVIntrinsicUtils.cpp
Go to the documentation of this file.
1//===- RISCVVIntrinsicUtils.cpp - RISC-V Vector Intrinsic Utils -*- C++ -*-===//
2//
3// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4// See https://llvm.org/LICENSE.txt for license information.
5// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6//
7//===----------------------------------------------------------------------===//
8
10#include "llvm/ADT/ArrayRef.h"
11#include "llvm/ADT/SmallSet.h"
12#include "llvm/ADT/StringExtras.h"
13#include "llvm/ADT/StringSet.h"
14#include "llvm/ADT/Twine.h"
15#include "llvm/Support/ErrorHandling.h"
16#include "llvm/Support/raw_ostream.h"
17#include <numeric>
18#include <optional>
19
20using namespace llvm;
21
22namespace clang {
23namespace RISCV {
24
25const PrototypeDescriptor PrototypeDescriptor::Mask = PrototypeDescriptor(
27const PrototypeDescriptor PrototypeDescriptor::VL =
28 PrototypeDescriptor(BaseTypeModifier::SizeT);
29const PrototypeDescriptor PrototypeDescriptor::Vector =
30 PrototypeDescriptor(BaseTypeModifier::Vector);
31
32//===----------------------------------------------------------------------===//
33// Type implementation
34//===----------------------------------------------------------------------===//
35
36LMULType::LMULType(int NewLog2LMUL) {
37 // Check Log2LMUL is -3, -2, -1, 0, 1, 2, 3
38 assert(NewLog2LMUL <= 3 && NewLog2LMUL >= -3 && "Bad LMUL number!");
39 Log2LMUL = NewLog2LMUL;
40}
41
42std::string LMULType::str() const {
43 if (Log2LMUL < 0)
44 return "mf" + utostr(1ULL << (-Log2LMUL));
45 return "m" + utostr(1ULL << Log2LMUL);
46}
47
48VScaleVal LMULType::getScale(unsigned ElementBitwidth) const {
49 int Log2ScaleResult = 0;
50 switch (ElementBitwidth) {
51 default:
52 break;
53 case 8:
54 Log2ScaleResult = Log2LMUL + 3;
55 break;
56 case 16:
57 Log2ScaleResult = Log2LMUL + 2;
58 break;
59 case 32:
60 Log2ScaleResult = Log2LMUL + 1;
61 break;
62 case 64:
63 Log2ScaleResult = Log2LMUL;
64 break;
65 }
66 // Illegal vscale result would be less than 1
67 if (Log2ScaleResult < 0)
68 return std::nullopt;
69 return 1 << Log2ScaleResult;
70}
71
72void LMULType::MulLog2LMUL(int log2LMUL) { Log2LMUL += log2LMUL; }
73
74RVVType::RVVType(BasicType BT, int Log2LMUL,
75 const PrototypeDescriptor &prototype)
76 : BT(BT), LMUL(LMULType(Log2LMUL)) {
77 applyBasicType();
78 applyModifier(prototype);
79 Valid = verifyType();
80 if (Valid) {
81 initBuiltinStr();
82 initTypeStr();
83 if (isVector()) {
84 initClangBuiltinStr();
85 }
86 }
87}
88
89// clang-format off
90// boolean type are encoded the ratio of n (SEW/LMUL)
91// SEW/LMUL | 1 | 2 | 4 | 8 | 16 | 32 | 64
92// c type | vbool64_t | vbool32_t | vbool16_t | vbool8_t | vbool4_t | vbool2_t | vbool1_t
93// IR type | nxv1i1 | nxv2i1 | nxv4i1 | nxv8i1 | nxv16i1 | nxv32i1 | nxv64i1
94
95// type\lmul | 1/8 | 1/4 | 1/2 | 1 | 2 | 4 | 8
96// -------- |------ | -------- | ------- | ------- | -------- | -------- | --------
97// i64 | N/A | N/A | N/A | nxv1i64 | nxv2i64 | nxv4i64 | nxv8i64
98// i32 | N/A | N/A | nxv1i32 | nxv2i32 | nxv4i32 | nxv8i32 | nxv16i32
99// i16 | N/A | nxv1i16 | nxv2i16 | nxv4i16 | nxv8i16 | nxv16i16 | nxv32i16
100// i8 | nxv1i8 | nxv2i8 | nxv4i8 | nxv8i8 | nxv16i8 | nxv32i8 | nxv64i8
101// double | N/A | N/A | N/A | nxv1f64 | nxv2f64 | nxv4f64 | nxv8f64
102// float | N/A | N/A | nxv1f32 | nxv2f32 | nxv4f32 | nxv8f32 | nxv16f32
103// half | N/A | nxv1f16 | nxv2f16 | nxv4f16 | nxv8f16 | nxv16f16 | nxv32f16
104// bfloat16 | N/A | nxv1bf16 | nxv2bf16| nxv4bf16| nxv8bf16 | nxv16bf16| nxv32bf16
105// clang-format on
106
107bool RVVType::verifyType() const {
108 if (ScalarType == Invalid)
109 return false;
110 if (isScalar())
111 return true;
112 if (!Scale)
113 return false;
114 if (isFloat() && ElementBitwidth == 8)
115 return false;
116 if (isBFloat() && ElementBitwidth != 16)
117 return false;
118 if (IsTuple && (NF == 1 || NF > 8))
119 return false;
120 if (IsTuple && (1 << std::max(0, LMUL.Log2LMUL)) * NF > 8)
121 return false;
122 unsigned V = *Scale;
123 switch (ElementBitwidth) {
124 case 1:
125 case 8:
126 // Check Scale is 1,2,4,8,16,32,64
127 return (V <= 64 && isPowerOf2_32(V));
128 case 16:
129 // Check Scale is 1,2,4,8,16,32
130 return (V <= 32 && isPowerOf2_32(V));
131 case 32:
132 // Check Scale is 1,2,4,8,16
133 return (V <= 16 && isPowerOf2_32(V));
134 case 64:
135 // Check Scale is 1,2,4,8
136 return (V <= 8 && isPowerOf2_32(V));
137 }
138 return false;
139}
140
141void RVVType::initBuiltinStr() {
142 assert(isValid() && "RVVType is invalid");
143 switch (ScalarType) {
145 BuiltinStr = "v";
146 return;
148 BuiltinStr = "z";
149 if (IsImmediate)
150 BuiltinStr = "I" + BuiltinStr;
151 if (IsPointer)
152 BuiltinStr += "*";
153 return;
155 BuiltinStr = "Y";
156 return;
158 BuiltinStr = "ULi";
159 return;
161 BuiltinStr = "Li";
162 return;
164 assert(ElementBitwidth == 1);
165 BuiltinStr += "b";
166 break;
169 switch (ElementBitwidth) {
170 case 8:
171 BuiltinStr += "c";
172 break;
173 case 16:
174 BuiltinStr += "s";
175 break;
176 case 32:
177 BuiltinStr += "i";
178 break;
179 case 64:
180 BuiltinStr += "Wi";
181 break;
182 default:
183 llvm_unreachable("Unhandled ElementBitwidth!");
184 }
185 if (isSignedInteger())
186 BuiltinStr = "S" + BuiltinStr;
187 else
188 BuiltinStr = "U" + BuiltinStr;
189 break;
191 switch (ElementBitwidth) {
192 case 16:
193 BuiltinStr += "x";
194 break;
195 case 32:
196 BuiltinStr += "f";
197 break;
198 case 64:
199 BuiltinStr += "d";
200 break;
201 default:
202 llvm_unreachable("Unhandled ElementBitwidth!");
203 }
204 break;
206 BuiltinStr += "y";
207 break;
208 default:
209 llvm_unreachable("ScalarType is invalid!");
210 }
211 if (IsImmediate)
212 BuiltinStr = "I" + BuiltinStr;
213 if (isScalar()) {
214 if (IsConstant)
215 BuiltinStr += "C";
216 if (IsPointer)
217 BuiltinStr += "*";
218 return;
219 }
220 BuiltinStr = "q" + utostr(*Scale) + BuiltinStr;
221 // Pointer to vector types. Defined for segment load intrinsics.
222 // segment load intrinsics have pointer type arguments to store the loaded
223 // vector values.
224 if (IsPointer)
225 BuiltinStr += "*";
226
227 if (IsTuple)
228 BuiltinStr = "T" + utostr(NF) + BuiltinStr;
229}
230
231void RVVType::initClangBuiltinStr() {
232 assert(isValid() && "RVVType is invalid");
233 assert(isVector() && "Handle Vector type only");
234
235 ClangBuiltinStr = "__rvv_";
236 switch (ScalarType) {
238 ClangBuiltinStr += "bool" + utostr(64 / *Scale) + "_t";
239 return;
241 ClangBuiltinStr += "float";
242 break;
244 ClangBuiltinStr += "bfloat";
245 break;
247 ClangBuiltinStr += "int";
248 break;
250 ClangBuiltinStr += "uint";
251 break;
252 default:
253 llvm_unreachable("ScalarTypeKind is invalid");
254 }
255 ClangBuiltinStr += utostr(ElementBitwidth) + LMUL.str() +
256 (IsTuple ? "x" + utostr(NF) : "") + "_t";
257}
258
259void RVVType::initTypeStr() {
260 assert(isValid() && "RVVType is invalid");
261
262 if (IsConstant)
263 Str += "const ";
264
265 auto getTypeString = [&](StringRef TypeStr) {
266 if (isScalar())
267 return Twine(TypeStr + Twine(ElementBitwidth) + "_t").str();
268 return Twine("v" + TypeStr + Twine(ElementBitwidth) + LMUL.str() +
269 (IsTuple ? "x" + utostr(NF) : "") + "_t")
270 .str();
271 };
272
273 switch (ScalarType) {
275 Str = "void";
276 return;
278 Str = "size_t";
279 if (IsPointer)
280 Str += " *";
281 return;
283 Str = "ptrdiff_t";
284 return;
286 Str = "unsigned long";
287 return;
289 Str = "long";
290 return;
292 if (isScalar())
293 Str += "bool";
294 else
295 // Vector bool is special case, the formulate is
296 // `vbool<N>_t = MVT::nxv<64/N>i1` ex. vbool16_t = MVT::4i1
297 Str += "vbool" + utostr(64 / *Scale) + "_t";
298 break;
300 if (isScalar()) {
301 if (ElementBitwidth == 64)
302 Str += "double";
303 else if (ElementBitwidth == 32)
304 Str += "float";
305 else if (ElementBitwidth == 16)
306 Str += "_Float16";
307 else
308 llvm_unreachable("Unhandled floating type.");
309 } else
310 Str += getTypeString("float");
311 break;
313 if (isScalar()) {
314 if (ElementBitwidth == 16)
315 Str += "__bf16";
316 else
317 llvm_unreachable("Unhandled floating type.");
318 } else
319 Str += getTypeString("bfloat");
320 break;
322 Str += getTypeString("int");
323 break;
325 Str += getTypeString("uint");
326 break;
327 default:
328 llvm_unreachable("ScalarType is invalid!");
329 }
330 if (IsPointer)
331 Str += " *";
332}
333
334void RVVType::initShortStr() {
335 switch (ScalarType) {
337 assert(isVector());
338 ShortStr = "b" + utostr(64 / *Scale);
339 return;
341 ShortStr = "f" + utostr(ElementBitwidth);
342 break;
344 ShortStr = "bf" + utostr(ElementBitwidth);
345 break;
347 ShortStr = "i" + utostr(ElementBitwidth);
348 break;
350 ShortStr = "u" + utostr(ElementBitwidth);
351 break;
352 default:
353 llvm_unreachable("Unhandled case!");
354 }
355 if (isVector())
356 ShortStr += LMUL.str();
357 if (isTuple())
358 ShortStr += "x" + utostr(NF);
359}
360
361static VectorTypeModifier getTupleVTM(unsigned NF) {
362 assert(2 <= NF && NF <= 8 && "2 <= NF <= 8");
363 return static_cast<VectorTypeModifier>(
364 static_cast<uint8_t>(VectorTypeModifier::Tuple2) + (NF - 2));
365}
366
367void RVVType::applyBasicType() {
368 switch (BT) {
369 case BasicType::Int8:
370 ElementBitwidth = 8;
372 break;
373 case BasicType::Int16:
374 ElementBitwidth = 16;
376 break;
377 case BasicType::Int32:
378 ElementBitwidth = 32;
380 break;
381 case BasicType::Int64:
382 ElementBitwidth = 64;
384 break;
386 ElementBitwidth = 16;
387 ScalarType = ScalarTypeKind::Float;
388 break;
390 ElementBitwidth = 32;
391 ScalarType = ScalarTypeKind::Float;
392 break;
394 ElementBitwidth = 64;
395 ScalarType = ScalarTypeKind::Float;
396 break;
398 ElementBitwidth = 16;
399 ScalarType = ScalarTypeKind::BFloat;
400 break;
401 default:
402 llvm_unreachable("Unhandled type code!");
403 }
404 assert(ElementBitwidth != 0 && "Bad element bitwidth!");
405}
406
407std::optional<PrototypeDescriptor>
409 llvm::StringRef PrototypeDescriptorStr) {
413
414 if (PrototypeDescriptorStr.empty())
415 return PD;
416
417 // Handle base type modifier
418 auto PType = PrototypeDescriptorStr.back();
419 switch (PType) {
420 case 'e':
422 break;
423 case 'v':
425 break;
426 case 'w':
429 break;
430 case 'q':
433 break;
434 case 'o':
437 break;
438 case 'm':
441 break;
442 case '0':
444 break;
445 case 'z':
447 break;
448 case 't':
450 break;
451 case 'u':
453 break;
454 case 'l':
456 break;
457 case 'f':
459 break;
460 default:
461 llvm_unreachable("Illegal primitive type transformers!");
462 }
463 PD.PT = static_cast<uint8_t>(PT);
464 PrototypeDescriptorStr = PrototypeDescriptorStr.drop_back();
465
466 // Compute the vector type transformers, it can only appear one time.
467 if (PrototypeDescriptorStr.starts_with("(")) {
469 "VectorTypeModifier should only have one modifier");
470 size_t Idx = PrototypeDescriptorStr.find(')');
471 assert(Idx != StringRef::npos);
472 StringRef ComplexType = PrototypeDescriptorStr.slice(1, Idx);
473 PrototypeDescriptorStr = PrototypeDescriptorStr.drop_front(Idx + 1);
474 assert(!PrototypeDescriptorStr.contains('(') &&
475 "Only allow one vector type modifier");
476
477 auto ComplexTT = ComplexType.split(":");
478 if (ComplexTT.first == "Log2EEW") {
479 uint32_t Log2EEW;
480 if (ComplexTT.second.getAsInteger(10, Log2EEW)) {
481 llvm_unreachable("Invalid Log2EEW value!");
482 return std::nullopt;
483 }
484 switch (Log2EEW) {
485 case 3:
487 break;
488 case 4:
490 break;
491 case 5:
493 break;
494 case 6:
496 break;
497 default:
498 llvm_unreachable("Invalid Log2EEW value, should be [3-6]");
499 return std::nullopt;
500 }
501 } else if (ComplexTT.first == "FixedSEW") {
502 uint32_t NewSEW;
503 if (ComplexTT.second.getAsInteger(10, NewSEW)) {
504 llvm_unreachable("Invalid FixedSEW value!");
505 return std::nullopt;
506 }
507 switch (NewSEW) {
508 case 8:
510 break;
511 case 16:
513 break;
514 case 32:
516 break;
517 case 64:
519 break;
520 default:
521 llvm_unreachable("Invalid FixedSEW value, should be 8, 16, 32 or 64");
522 return std::nullopt;
523 }
524 } else if (ComplexTT.first == "LFixedLog2LMUL") {
525 int32_t Log2LMUL;
526 if (ComplexTT.second.getAsInteger(10, Log2LMUL)) {
527 llvm_unreachable("Invalid LFixedLog2LMUL value!");
528 return std::nullopt;
529 }
530 switch (Log2LMUL) {
531 case -3:
533 break;
534 case -2:
536 break;
537 case -1:
539 break;
540 case 0:
542 break;
543 case 1:
545 break;
546 case 2:
548 break;
549 case 3:
551 break;
552 default:
553 llvm_unreachable("Invalid LFixedLog2LMUL value, should be [-3, 3]");
554 return std::nullopt;
555 }
556 } else if (ComplexTT.first == "SFixedLog2LMUL") {
557 int32_t Log2LMUL;
558 if (ComplexTT.second.getAsInteger(10, Log2LMUL)) {
559 llvm_unreachable("Invalid SFixedLog2LMUL value!");
560 return std::nullopt;
561 }
562 switch (Log2LMUL) {
563 case -3:
565 break;
566 case -2:
568 break;
569 case -1:
571 break;
572 case 0:
574 break;
575 case 1:
577 break;
578 case 2:
580 break;
581 case 3:
583 break;
584 default:
585 llvm_unreachable("Invalid LFixedLog2LMUL value, should be [-3, 3]");
586 return std::nullopt;
587 }
588
589 } else if (ComplexTT.first == "SEFixedLog2LMUL") {
590 int32_t Log2LMUL;
591 if (ComplexTT.second.getAsInteger(10, Log2LMUL)) {
592 llvm_unreachable("Invalid SEFixedLog2LMUL value!");
593 return std::nullopt;
594 }
595 switch (Log2LMUL) {
596 case -3:
598 break;
599 case -2:
601 break;
602 case -1:
604 break;
605 case 0:
607 break;
608 case 1:
610 break;
611 case 2:
613 break;
614 case 3:
616 break;
617 default:
618 llvm_unreachable("Invalid LFixedLog2LMUL value, should be [-3, 3]");
619 return std::nullopt;
620 }
621 } else if (ComplexTT.first == "Tuple") {
622 unsigned NF = 0;
623 if (ComplexTT.second.getAsInteger(10, NF)) {
624 llvm_unreachable("Invalid NF value!");
625 return std::nullopt;
626 }
627 VTM = getTupleVTM(NF);
628 } else {
629 llvm_unreachable("Illegal complex type transformers!");
630 }
631 }
632 PD.VTM = static_cast<uint8_t>(VTM);
633
634 // Compute the remain type transformers
636 for (char I : PrototypeDescriptorStr) {
637 switch (I) {
638 case 'P':
640 llvm_unreachable("'P' transformer cannot be used after 'C'");
642 llvm_unreachable("'P' transformer cannot be used twice");
644 break;
645 case 'C':
647 break;
648 case 'K':
650 break;
651 case 'U':
653 break;
654 case 'I':
656 break;
657 case 'F':
659 break;
660 case 'S':
662 break;
663 default:
664 llvm_unreachable("Illegal non-primitive type transformer!");
665 }
666 }
667 PD.TM = static_cast<uint8_t>(TM);
668
669 return PD;
670}
671
672void RVVType::applyModifier(const PrototypeDescriptor &Transformer) {
673 // Handle primitive type transformer
674 switch (static_cast<BaseTypeModifier>(Transformer.PT)) {
676 Scale = 0;
677 break;
679 Scale = LMUL.getScale(ElementBitwidth);
680 break;
682 ScalarType = ScalarTypeKind::Void;
683 break;
685 ScalarType = ScalarTypeKind::Size_t;
686 break;
688 ScalarType = ScalarTypeKind::Ptrdiff_t;
689 break;
691 ScalarType = ScalarTypeKind::UnsignedLong;
692 break;
694 ScalarType = ScalarTypeKind::SignedLong;
695 break;
697 ElementBitwidth = 32;
698 ScalarType = ScalarTypeKind::Float;
699 break;
701 ScalarType = ScalarTypeKind::Invalid;
702 return;
703 }
704
705 switch (static_cast<VectorTypeModifier>(Transformer.VTM)) {
707 ElementBitwidth *= 2;
708 LMUL.MulLog2LMUL(1);
709 Scale = LMUL.getScale(ElementBitwidth);
710 break;
712 ElementBitwidth *= 4;
713 LMUL.MulLog2LMUL(2);
714 Scale = LMUL.getScale(ElementBitwidth);
715 break;
717 ElementBitwidth *= 8;
718 LMUL.MulLog2LMUL(3);
719 Scale = LMUL.getScale(ElementBitwidth);
720 break;
722 ScalarType = ScalarTypeKind::Boolean;
723 Scale = LMUL.getScale(ElementBitwidth);
724 ElementBitwidth = 1;
725 break;
727 applyLog2EEW(3);
728 break;
730 applyLog2EEW(4);
731 break;
733 applyLog2EEW(5);
734 break;
736 applyLog2EEW(6);
737 break;
739 applyFixedSEW(8);
740 break;
742 applyFixedSEW(16);
743 break;
745 applyFixedSEW(32);
746 break;
748 applyFixedSEW(64);
749 break;
751 applyFixedLog2LMUL(-3, FixedLMULType::LargerThan);
752 break;
754 applyFixedLog2LMUL(-2, FixedLMULType::LargerThan);
755 break;
757 applyFixedLog2LMUL(-1, FixedLMULType::LargerThan);
758 break;
760 applyFixedLog2LMUL(0, FixedLMULType::LargerThan);
761 break;
763 applyFixedLog2LMUL(1, FixedLMULType::LargerThan);
764 break;
766 applyFixedLog2LMUL(2, FixedLMULType::LargerThan);
767 break;
769 applyFixedLog2LMUL(3, FixedLMULType::LargerThan);
770 break;
772 applyFixedLog2LMUL(-3, FixedLMULType::SmallerThan);
773 break;
775 applyFixedLog2LMUL(-2, FixedLMULType::SmallerThan);
776 break;
778 applyFixedLog2LMUL(-1, FixedLMULType::SmallerThan);
779 break;
781 applyFixedLog2LMUL(0, FixedLMULType::SmallerThan);
782 break;
784 applyFixedLog2LMUL(1, FixedLMULType::SmallerThan);
785 break;
787 applyFixedLog2LMUL(2, FixedLMULType::SmallerThan);
788 break;
790 applyFixedLog2LMUL(3, FixedLMULType::SmallerThan);
791 break;
793 applyFixedLog2LMUL(-3, FixedLMULType::SmallerOrEqual);
794 break;
796 applyFixedLog2LMUL(-2, FixedLMULType::SmallerOrEqual);
797 break;
799 applyFixedLog2LMUL(-1, FixedLMULType::SmallerOrEqual);
800 break;
802 applyFixedLog2LMUL(0, FixedLMULType::SmallerOrEqual);
803 break;
805 applyFixedLog2LMUL(1, FixedLMULType::SmallerOrEqual);
806 break;
808 applyFixedLog2LMUL(2, FixedLMULType::SmallerOrEqual);
809 break;
811 applyFixedLog2LMUL(3, FixedLMULType::SmallerOrEqual);
812 break;
820 IsTuple = true;
821 NF = 2 + static_cast<uint8_t>(Transformer.VTM) -
822 static_cast<uint8_t>(VectorTypeModifier::Tuple2);
823 break;
824 }
826 break;
827 }
828
829 // Early return if the current type modifier is already invalid.
830 if (ScalarType == Invalid)
831 return;
832
833 for (unsigned TypeModifierMaskShift = 0;
834 TypeModifierMaskShift <= static_cast<unsigned>(TypeModifier::MaxOffset);
835 ++TypeModifierMaskShift) {
836 unsigned TypeModifierMask = 1 << TypeModifierMaskShift;
837 if ((static_cast<unsigned>(Transformer.TM) & TypeModifierMask) !=
838 TypeModifierMask)
839 continue;
840 switch (static_cast<TypeModifier>(TypeModifierMask)) {
842 IsPointer = true;
843 break;
845 IsConstant = true;
846 break;
848 IsImmediate = true;
849 IsConstant = true;
850 break;
853 break;
856 break;
858 ScalarType = ScalarTypeKind::Float;
859 break;
861 ScalarType = ScalarTypeKind::BFloat;
862 break;
864 LMUL = LMULType(0);
865 // Update ElementBitwidth need to update Scale too.
866 Scale = LMUL.getScale(ElementBitwidth);
867 break;
868 default:
869 llvm_unreachable("Unknown type modifier mask!");
870 }
871 }
872}
873
874void RVVType::applyLog2EEW(unsigned Log2EEW) {
875 // update new elmul = (eew/sew) * lmul
876 LMUL.MulLog2LMUL(Log2EEW - Log2_32(ElementBitwidth));
877 // update new eew
878 ElementBitwidth = 1 << Log2EEW;
880 Scale = LMUL.getScale(ElementBitwidth);
881}
882
883void RVVType::applyFixedSEW(unsigned NewSEW) {
884 // Set invalid type if src and dst SEW are same.
885 if (ElementBitwidth == NewSEW) {
886 ScalarType = ScalarTypeKind::Invalid;
887 return;
888 }
889 // Update new SEW
890 ElementBitwidth = NewSEW;
891 Scale = LMUL.getScale(ElementBitwidth);
892}
893
894void RVVType::applyFixedLog2LMUL(int Log2LMUL, enum FixedLMULType Type) {
895 switch (Type) {
896 case FixedLMULType::LargerThan:
897 if (Log2LMUL <= LMUL.Log2LMUL) {
898 ScalarType = ScalarTypeKind::Invalid;
899 return;
900 }
901 break;
902 case FixedLMULType::SmallerThan:
903 if (Log2LMUL >= LMUL.Log2LMUL) {
904 ScalarType = ScalarTypeKind::Invalid;
905 return;
906 }
907 break;
908 case FixedLMULType::SmallerOrEqual:
909 if (Log2LMUL > LMUL.Log2LMUL) {
910 ScalarType = ScalarTypeKind::Invalid;
911 return;
912 }
913 break;
914 }
915
916 // Update new LMUL
917 LMUL = LMULType(Log2LMUL);
918 Scale = LMUL.getScale(ElementBitwidth);
919}
920
921std::optional<RVVTypes>
922RVVTypeCache::computeTypes(BasicType BT, int Log2LMUL, unsigned NF,
924 RVVTypes Types;
925 for (const PrototypeDescriptor &Proto : Prototype) {
926 auto T = computeType(BT, Log2LMUL, Proto);
927 if (!T)
928 return std::nullopt;
929 // Record legal type index
930 Types.push_back(*T);
931 }
932 return Types;
933}
934
935// Compute the hash value of RVVType, used for cache the result of computeType.
936static uint64_t computeRVVTypeHashValue(BasicType BT, int Log2LMUL,
937 PrototypeDescriptor Proto) {
938 // Layout of hash value:
939 // 0 8 16 24 32 40
940 // | Log2LMUL + 3 | BT | Proto.PT | Proto.TM | Proto.VTM |
941 assert(Log2LMUL >= -3 && Log2LMUL <= 3);
942 return (Log2LMUL + 3) | (static_cast<uint64_t>(BT) & 0xff) << 8 |
943 ((uint64_t)(Proto.PT & 0xff) << 16) |
944 ((uint64_t)(Proto.TM & 0xff) << 24) |
945 ((uint64_t)(Proto.VTM & 0xff) << 32);
946}
947
948std::optional<RVVTypePtr> RVVTypeCache::computeType(BasicType BT, int Log2LMUL,
949 PrototypeDescriptor Proto) {
950 uint64_t Idx = computeRVVTypeHashValue(BT, Log2LMUL, Proto);
951 // Search first
952 auto It = LegalTypes.find(Idx);
953 if (It != LegalTypes.end())
954 return &(It->second);
955
956 if (IllegalTypes.count(Idx))
957 return std::nullopt;
958
959 // Compute type and record the result.
960 RVVType T(BT, Log2LMUL, Proto);
961 if (T.isValid()) {
962 // Record legal type index and value.
963 std::pair<std::unordered_map<uint64_t, RVVType>::iterator, bool>
964 InsertResult = LegalTypes.insert({Idx, T});
965 return &(InsertResult.first->second);
966 }
967 // Record illegal type index.
968 IllegalTypes.insert(Idx);
969 return std::nullopt;
970}
971
972//===----------------------------------------------------------------------===//
973// RVVIntrinsic implementation
974//===----------------------------------------------------------------------===//
976 StringRef NewName, StringRef Suffix, StringRef NewOverloadedName,
977 StringRef OverloadedSuffix, StringRef IRName, bool IsMasked,
978 bool HasMaskedOffOperand, bool HasVL, PolicyScheme Scheme,
979 bool SupportOverloading, bool HasBuiltinAlias, StringRef ManualCodegen,
980 const RVVTypes &OutInTypes, const std::vector<int64_t> &NewIntrinsicTypes,
981 const std::vector<StringRef> &RequiredFeatures, unsigned NF,
982 Policy NewPolicyAttrs, bool HasFRMRoundModeOp)
983 : IRName(IRName), IsMasked(IsMasked),
984 HasMaskedOffOperand(HasMaskedOffOperand), HasVL(HasVL), Scheme(Scheme),
985 SupportOverloading(SupportOverloading), HasBuiltinAlias(HasBuiltinAlias),
986 ManualCodegen(ManualCodegen.str()), NF(NF), PolicyAttrs(NewPolicyAttrs) {
987
988 // Init BuiltinName, Name and OverloadedName
989 BuiltinName = NewName.str();
990 Name = BuiltinName;
991 if (NewOverloadedName.empty())
992 OverloadedName = NewName.split("_").first.str();
993 else
994 OverloadedName = NewOverloadedName.str();
995 if (!Suffix.empty())
996 Name += "_" + Suffix.str();
997 if (!OverloadedSuffix.empty())
998 OverloadedName += "_" + OverloadedSuffix.str();
999
1000 updateNamesAndPolicy(IsMasked, hasPolicy(), Name, BuiltinName, OverloadedName,
1001 PolicyAttrs, HasFRMRoundModeOp);
1002
1003 // Init OutputType and InputTypes
1004 OutputType = OutInTypes[0];
1005 InputTypes.assign(OutInTypes.begin() + 1, OutInTypes.end());
1006
1007 // IntrinsicTypes is unmasked TA version index. Need to update it
1008 // if there is merge operand (It is always in first operand).
1009 IntrinsicTypes = NewIntrinsicTypes;
1010 if ((IsMasked && hasMaskedOffOperand()) ||
1011 (!IsMasked && hasPassthruOperand())) {
1012 for (auto &I : IntrinsicTypes) {
1013 if (I >= 0)
1014 I += NF;
1015 }
1016 }
1017}
1018
1020 std::string S;
1021 S += OutputType->getBuiltinStr();
1022 for (const auto &T : InputTypes) {
1023 S += T->getBuiltinStr();
1024 }
1025 return S;
1026}
1027
1029 RVVTypeCache &TypeCache, BasicType Type, int Log2LMUL,
1030 llvm::ArrayRef<PrototypeDescriptor> PrototypeDescriptors) {
1031 SmallVector<std::string> SuffixStrs;
1032 for (auto PD : PrototypeDescriptors) {
1033 auto T = TypeCache.computeType(Type, Log2LMUL, PD);
1034 SuffixStrs.push_back((*T)->getShortStr());
1035 }
1036 return join(SuffixStrs, "_");
1037}
1038
1041 bool HasMaskedOffOperand, bool HasVL, unsigned NF,
1042 PolicyScheme DefaultScheme, Policy PolicyAttrs, bool IsTuple) {
1043 SmallVector<PrototypeDescriptor> NewPrototype(Prototype.begin(),
1044 Prototype.end());
1045 bool HasPassthruOp = DefaultScheme == PolicyScheme::HasPassthruOperand;
1046 if (IsMasked) {
1047 // If HasMaskedOffOperand, insert result type as first input operand if
1048 // need.
1049 if (HasMaskedOffOperand && !PolicyAttrs.isTAMAPolicy()) {
1050 if (NF == 1) {
1051 NewPrototype.insert(NewPrototype.begin() + 1, NewPrototype[0]);
1052 } else if (NF > 1) {
1053 if (IsTuple) {
1054 PrototypeDescriptor BasePtrOperand = Prototype[1];
1056 static_cast<uint8_t>(BaseTypeModifier::Vector),
1057 static_cast<uint8_t>(getTupleVTM(NF)),
1058 BasePtrOperand.TM & ~static_cast<uint8_t>(TypeModifier::Pointer));
1059 NewPrototype.insert(NewPrototype.begin() + 1, MaskoffType);
1060 } else {
1061 // Convert
1062 // (void, op0 address, op1 address, ...)
1063 // to
1064 // (void, op0 address, op1 address, ..., maskedoff0, maskedoff1, ...)
1065 PrototypeDescriptor MaskoffType = NewPrototype[1];
1066 MaskoffType.TM &= ~static_cast<uint8_t>(TypeModifier::Pointer);
1067 NewPrototype.insert(NewPrototype.begin() + NF + 1, NF, MaskoffType);
1068 }
1069 }
1070 }
1071 if (HasMaskedOffOperand && NF > 1) {
1072 // Convert
1073 // (void, op0 address, op1 address, ..., maskedoff0, maskedoff1, ...)
1074 // to
1075 // (void, op0 address, op1 address, ..., mask, maskedoff0, maskedoff1,
1076 // ...)
1077 if (IsTuple)
1078 NewPrototype.insert(NewPrototype.begin() + 1,
1080 else
1081 NewPrototype.insert(NewPrototype.begin() + NF + 1,
1083 } else {
1084 // If IsMasked, insert PrototypeDescriptor:Mask as first input operand.
1085 NewPrototype.insert(NewPrototype.begin() + 1, PrototypeDescriptor::Mask);
1086 }
1087 } else {
1088 if (NF == 1) {
1089 if (PolicyAttrs.isTUPolicy() && HasPassthruOp)
1090 NewPrototype.insert(NewPrototype.begin(), NewPrototype[0]);
1091 } else if (PolicyAttrs.isTUPolicy() && HasPassthruOp) {
1092 if (IsTuple) {
1093 PrototypeDescriptor BasePtrOperand = Prototype[0];
1095 static_cast<uint8_t>(BaseTypeModifier::Vector),
1096 static_cast<uint8_t>(getTupleVTM(NF)),
1097 BasePtrOperand.TM & ~static_cast<uint8_t>(TypeModifier::Pointer));
1098 NewPrototype.insert(NewPrototype.begin(), MaskoffType);
1099 } else {
1100 // NF > 1 cases for segment load operations.
1101 // Convert
1102 // (void, op0 address, op1 address, ...)
1103 // to
1104 // (void, op0 address, op1 address, maskedoff0, maskedoff1, ...)
1105 PrototypeDescriptor MaskoffType = Prototype[1];
1106 MaskoffType.TM &= ~static_cast<uint8_t>(TypeModifier::Pointer);
1107 NewPrototype.insert(NewPrototype.begin() + NF + 1, NF, MaskoffType);
1108 }
1109 }
1110 }
1111
1112 // If HasVL, append PrototypeDescriptor:VL to last operand
1113 if (HasVL)
1114 NewPrototype.push_back(PrototypeDescriptor::VL);
1115
1116 return NewPrototype;
1117}
1118
1121}
1122
1125 bool HasMaskPolicy) {
1126 if (HasTailPolicy && HasMaskPolicy)
1133 if (HasTailPolicy && !HasMaskPolicy)
1136 if (!HasTailPolicy && HasMaskPolicy)
1139 llvm_unreachable("An RVV instruction should not be without both tail policy "
1140 "and mask policy");
1141}
1142
1144 bool IsMasked, bool HasPolicy, std::string &Name, std::string &BuiltinName,
1145 std::string &OverloadedName, Policy &PolicyAttrs, bool HasFRMRoundModeOp) {
1146
1147 auto appendPolicySuffix = [&](const std::string &suffix) {
1148 Name += suffix;
1149 BuiltinName += suffix;
1150 OverloadedName += suffix;
1151 };
1152
1153 if (HasFRMRoundModeOp) {
1154 Name += "_rm";
1155 BuiltinName += "_rm";
1156 }
1157
1158 if (IsMasked) {
1159 if (PolicyAttrs.isTUMUPolicy())
1160 appendPolicySuffix("_tumu");
1161 else if (PolicyAttrs.isTUMAPolicy())
1162 appendPolicySuffix("_tum");
1163 else if (PolicyAttrs.isTAMUPolicy())
1164 appendPolicySuffix("_mu");
1165 else if (PolicyAttrs.isTAMAPolicy()) {
1166 Name += "_m";
1167 BuiltinName += "_m";
1168 } else
1169 llvm_unreachable("Unhandled policy condition");
1170 } else {
1171 if (PolicyAttrs.isTUPolicy())
1172 appendPolicySuffix("_tu");
1173 else if (PolicyAttrs.isTAPolicy()) // no suffix needed
1174 return;
1175 else
1176 llvm_unreachable("Unhandled policy condition");
1177 }
1178}
1179
1181 SmallVector<PrototypeDescriptor> PrototypeDescriptors;
1182 const StringRef Primaries("evwqom0ztulf");
1183 while (!Prototypes.empty()) {
1184 size_t Idx = 0;
1185 // Skip over complex prototype because it could contain primitive type
1186 // character.
1187 if (Prototypes[0] == '(')
1188 Idx = Prototypes.find_first_of(')');
1189 Idx = Prototypes.find_first_of(Primaries, Idx);
1190 assert(Idx != StringRef::npos);
1192 Prototypes.slice(0, Idx + 1));
1193 if (!PD)
1194 llvm_unreachable("Error during parsing prototype.");
1195 PrototypeDescriptors.push_back(*PD);
1196 Prototypes = Prototypes.drop_front(Idx + 1);
1197 }
1198 return PrototypeDescriptors;
1199}
1200
1201raw_ostream &operator<<(raw_ostream &OS, const RVVIntrinsicRecord &Record) {
1202 OS << "{";
1203 OS << "\"" << Record.Name << "\",";
1204 if (Record.OverloadedName == nullptr ||
1205 StringRef(Record.OverloadedName).empty())
1206 OS << "nullptr,";
1207 else
1208 OS << "\"" << Record.OverloadedName << "\",";
1209 OS << Record.PrototypeIndex << ",";
1210 OS << Record.SuffixIndex << ",";
1211 OS << Record.OverloadedSuffixIndex << ",";
1212 OS << (int)Record.PrototypeLength << ",";
1213 OS << (int)Record.SuffixLength << ",";
1214 OS << (int)Record.OverloadedSuffixSize << ",";
1215 OS << Record.RequiredExtensions << ",";
1216 OS << (int)Record.TypeRangeMask << ",";
1217 OS << (int)Record.Log2LMULMask << ",";
1218 OS << (int)Record.NF << ",";
1219 OS << (int)Record.HasMasked << ",";
1220 OS << (int)Record.HasVL << ",";
1221 OS << (int)Record.HasMaskedOffOperand << ",";
1222 OS << (int)Record.HasTailPolicy << ",";
1223 OS << (int)Record.HasMaskPolicy << ",";
1224 OS << (int)Record.HasFRMRoundModeOp << ",";
1225 OS << (int)Record.IsTuple << ",";
1226 OS << (int)Record.UnMaskedPolicyScheme << ",";
1227 OS << (int)Record.MaskedPolicyScheme << ",";
1228 OS << "},\n";
1229 return OS;
1230}
1231
1232} // end namespace RISCV
1233} // end namespace clang
#define V(N, I)
Definition: ASTContext.h:3266
static bool getTypeString(SmallStringEnc &Enc, const Decl *D, const CodeGen::CodeGenModule &CGM, TypeStringCache &TSC)
The XCore ABI includes a type information section that communicates symbol type information to the li...
Definition: XCore.cpp:632
llvm::MachO::Record Record
Definition: MachO.h:28
static bool isVector(QualType QT, QualType ElementType)
This helper function returns true if QT is a vector type that has element type ElementType.
Definition: SemaExpr.cpp:10126
__device__ int
Complex values, per C99 6.2.5p11.
Definition: Type.h:2875
static llvm::SmallVector< Policy > getSupportedMaskedPolicies(bool HasTailPolicy, bool HasMaskPolicy)
static llvm::SmallVector< PrototypeDescriptor > computeBuiltinTypes(llvm::ArrayRef< PrototypeDescriptor > Prototype, bool IsMasked, bool HasMaskedOffOperand, bool HasVL, unsigned NF, PolicyScheme DefaultScheme, Policy PolicyAttrs, bool IsTuple)
static void updateNamesAndPolicy(bool IsMasked, bool HasPolicy, std::string &Name, std::string &BuiltinName, std::string &OverloadedName, Policy &PolicyAttrs, bool HasFRMRoundModeOp)
static std::string getSuffixStr(RVVTypeCache &TypeCache, BasicType Type, int Log2LMUL, llvm::ArrayRef< PrototypeDescriptor > PrototypeDescriptors)
RVVIntrinsic(llvm::StringRef Name, llvm::StringRef Suffix, llvm::StringRef OverloadedName, llvm::StringRef OverloadedSuffix, llvm::StringRef IRName, bool IsMasked, bool HasMaskedOffOperand, bool HasVL, PolicyScheme Scheme, bool SupportOverloading, bool HasBuiltinAlias, llvm::StringRef ManualCodegen, const RVVTypes &Types, const std::vector< int64_t > &IntrinsicTypes, const std::vector< llvm::StringRef > &RequiredFeatures, unsigned NF, Policy PolicyAttrs, bool HasFRMRoundModeOp)
static llvm::SmallVector< Policy > getSupportedUnMaskedPolicies()
std::optional< RVVTypePtr > computeType(BasicType BT, int Log2LMUL, PrototypeDescriptor Proto)
std::optional< RVVTypes > computeTypes(BasicType BT, int Log2LMUL, unsigned NF, llvm::ArrayRef< PrototypeDescriptor > Prototype)
Compute output and input types by applying different config (basic type and LMUL with type transforme...
const std::string & getBuiltinStr() const
The base class of the type hierarchy.
Definition: Type.h:1607
llvm::raw_ostream & operator<<(llvm::raw_ostream &OS, const RVVIntrinsicRecord &RVVInstrRecord)
llvm::SmallVector< PrototypeDescriptor > parsePrototypes(llvm::StringRef Prototypes)
static uint64_t computeRVVTypeHashValue(BasicType BT, int Log2LMUL, PrototypeDescriptor Proto)
std::optional< unsigned > VScaleVal
static VectorTypeModifier getTupleVTM(unsigned NF)
std::vector< RVVTypePtr > RVVTypes
The JSON file list parser is used to communicate input to InstallAPI.
YAML serialization mapping.
Definition: Dominators.h:30
std::optional< unsigned > getScale(unsigned ElementBitwidth) const
void MulLog2LMUL(int Log2LMUL)
static std::optional< PrototypeDescriptor > parsePrototypeDescriptor(llvm::StringRef PrototypeStr)
static const PrototypeDescriptor VL
static const PrototypeDescriptor Mask
static const PrototypeDescriptor Vector