clang 20.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 unsigned NF, Policy NewPolicyAttrs, bool HasFRMRoundModeOp)
982 : IRName(IRName), IsMasked(IsMasked),
983 HasMaskedOffOperand(HasMaskedOffOperand), HasVL(HasVL), Scheme(Scheme),
984 SupportOverloading(SupportOverloading), HasBuiltinAlias(HasBuiltinAlias),
985 ManualCodegen(ManualCodegen.str()), NF(NF), PolicyAttrs(NewPolicyAttrs) {
986
987 // Init BuiltinName, Name and OverloadedName
988 BuiltinName = NewName.str();
989 Name = BuiltinName;
990 if (NewOverloadedName.empty())
991 OverloadedName = NewName.split("_").first.str();
992 else
993 OverloadedName = NewOverloadedName.str();
994 if (!Suffix.empty())
995 Name += "_" + Suffix.str();
996 if (!OverloadedSuffix.empty())
997 OverloadedName += "_" + OverloadedSuffix.str();
998
999 updateNamesAndPolicy(IsMasked, hasPolicy(), Name, BuiltinName, OverloadedName,
1000 PolicyAttrs, HasFRMRoundModeOp);
1001
1002 // Init OutputType and InputTypes
1003 OutputType = OutInTypes[0];
1004 InputTypes.assign(OutInTypes.begin() + 1, OutInTypes.end());
1005
1006 // IntrinsicTypes is unmasked TA version index. Need to update it
1007 // if there is merge operand (It is always in first operand).
1008 IntrinsicTypes = NewIntrinsicTypes;
1009 if ((IsMasked && hasMaskedOffOperand()) ||
1010 (!IsMasked && hasPassthruOperand())) {
1011 for (auto &I : IntrinsicTypes) {
1012 if (I >= 0)
1013 I += NF;
1014 }
1015 }
1016}
1017
1019 std::string S;
1020 S += OutputType->getBuiltinStr();
1021 for (const auto &T : InputTypes) {
1022 S += T->getBuiltinStr();
1023 }
1024 return S;
1025}
1026
1028 RVVTypeCache &TypeCache, BasicType Type, int Log2LMUL,
1029 llvm::ArrayRef<PrototypeDescriptor> PrototypeDescriptors) {
1030 SmallVector<std::string> SuffixStrs;
1031 for (auto PD : PrototypeDescriptors) {
1032 auto T = TypeCache.computeType(Type, Log2LMUL, PD);
1033 SuffixStrs.push_back((*T)->getShortStr());
1034 }
1035 return join(SuffixStrs, "_");
1036}
1037
1040 bool HasMaskedOffOperand, bool HasVL, unsigned NF,
1041 PolicyScheme DefaultScheme, Policy PolicyAttrs, bool IsTuple) {
1043 bool HasPassthruOp = DefaultScheme == PolicyScheme::HasPassthruOperand;
1044 if (IsMasked) {
1045 // If HasMaskedOffOperand, insert result type as first input operand if
1046 // need.
1047 if (HasMaskedOffOperand && !PolicyAttrs.isTAMAPolicy()) {
1048 if (NF == 1) {
1049 NewPrototype.insert(NewPrototype.begin() + 1, NewPrototype[0]);
1050 } else if (NF > 1) {
1051 if (IsTuple) {
1052 PrototypeDescriptor BasePtrOperand = Prototype[1];
1054 static_cast<uint8_t>(BaseTypeModifier::Vector),
1055 static_cast<uint8_t>(getTupleVTM(NF)),
1056 BasePtrOperand.TM & ~static_cast<uint8_t>(TypeModifier::Pointer));
1057 NewPrototype.insert(NewPrototype.begin() + 1, MaskoffType);
1058 } else {
1059 // Convert
1060 // (void, op0 address, op1 address, ...)
1061 // to
1062 // (void, op0 address, op1 address, ..., maskedoff0, maskedoff1, ...)
1063 PrototypeDescriptor MaskoffType = NewPrototype[1];
1064 MaskoffType.TM &= ~static_cast<uint8_t>(TypeModifier::Pointer);
1065 NewPrototype.insert(NewPrototype.begin() + NF + 1, NF, MaskoffType);
1066 }
1067 }
1068 }
1069 if (HasMaskedOffOperand && NF > 1) {
1070 // Convert
1071 // (void, op0 address, op1 address, ..., maskedoff0, maskedoff1, ...)
1072 // to
1073 // (void, op0 address, op1 address, ..., mask, maskedoff0, maskedoff1,
1074 // ...)
1075 if (IsTuple)
1076 NewPrototype.insert(NewPrototype.begin() + 1,
1078 else
1079 NewPrototype.insert(NewPrototype.begin() + NF + 1,
1081 } else {
1082 // If IsMasked, insert PrototypeDescriptor:Mask as first input operand.
1083 NewPrototype.insert(NewPrototype.begin() + 1, PrototypeDescriptor::Mask);
1084 }
1085 } else {
1086 if (NF == 1) {
1087 if (PolicyAttrs.isTUPolicy() && HasPassthruOp)
1088 NewPrototype.insert(NewPrototype.begin(), NewPrototype[0]);
1089 } else if (PolicyAttrs.isTUPolicy() && HasPassthruOp) {
1090 if (IsTuple) {
1091 PrototypeDescriptor BasePtrOperand = Prototype[0];
1093 static_cast<uint8_t>(BaseTypeModifier::Vector),
1094 static_cast<uint8_t>(getTupleVTM(NF)),
1095 BasePtrOperand.TM & ~static_cast<uint8_t>(TypeModifier::Pointer));
1096 NewPrototype.insert(NewPrototype.begin(), MaskoffType);
1097 } else {
1098 // NF > 1 cases for segment load operations.
1099 // Convert
1100 // (void, op0 address, op1 address, ...)
1101 // to
1102 // (void, op0 address, op1 address, maskedoff0, maskedoff1, ...)
1103 PrototypeDescriptor MaskoffType = Prototype[1];
1104 MaskoffType.TM &= ~static_cast<uint8_t>(TypeModifier::Pointer);
1105 NewPrototype.insert(NewPrototype.begin() + NF + 1, NF, MaskoffType);
1106 }
1107 }
1108 }
1109
1110 // If HasVL, append PrototypeDescriptor:VL to last operand
1111 if (HasVL)
1112 NewPrototype.push_back(PrototypeDescriptor::VL);
1113
1114 return NewPrototype;
1115}
1116
1119}
1120
1123 bool HasMaskPolicy) {
1124 if (HasTailPolicy && HasMaskPolicy)
1131 if (HasTailPolicy && !HasMaskPolicy)
1134 if (!HasTailPolicy && HasMaskPolicy)
1137 llvm_unreachable("An RVV instruction should not be without both tail policy "
1138 "and mask policy");
1139}
1140
1142 bool IsMasked, bool HasPolicy, std::string &Name, std::string &BuiltinName,
1143 std::string &OverloadedName, Policy &PolicyAttrs, bool HasFRMRoundModeOp) {
1144
1145 auto appendPolicySuffix = [&](const std::string &suffix) {
1146 Name += suffix;
1147 BuiltinName += suffix;
1148 OverloadedName += suffix;
1149 };
1150
1151 if (HasFRMRoundModeOp) {
1152 Name += "_rm";
1153 BuiltinName += "_rm";
1154 }
1155
1156 if (IsMasked) {
1157 if (PolicyAttrs.isTUMUPolicy())
1158 appendPolicySuffix("_tumu");
1159 else if (PolicyAttrs.isTUMAPolicy())
1160 appendPolicySuffix("_tum");
1161 else if (PolicyAttrs.isTAMUPolicy())
1162 appendPolicySuffix("_mu");
1163 else if (PolicyAttrs.isTAMAPolicy()) {
1164 Name += "_m";
1165 BuiltinName += "_m";
1166 } else
1167 llvm_unreachable("Unhandled policy condition");
1168 } else {
1169 if (PolicyAttrs.isTUPolicy())
1170 appendPolicySuffix("_tu");
1171 else if (PolicyAttrs.isTAPolicy()) // no suffix needed
1172 return;
1173 else
1174 llvm_unreachable("Unhandled policy condition");
1175 }
1176}
1177
1179 SmallVector<PrototypeDescriptor> PrototypeDescriptors;
1180 const StringRef Primaries("evwqom0ztulf");
1181 while (!Prototypes.empty()) {
1182 size_t Idx = 0;
1183 // Skip over complex prototype because it could contain primitive type
1184 // character.
1185 if (Prototypes[0] == '(')
1186 Idx = Prototypes.find_first_of(')');
1187 Idx = Prototypes.find_first_of(Primaries, Idx);
1188 assert(Idx != StringRef::npos);
1190 Prototypes.slice(0, Idx + 1));
1191 if (!PD)
1192 llvm_unreachable("Error during parsing prototype.");
1193 PrototypeDescriptors.push_back(*PD);
1194 Prototypes = Prototypes.drop_front(Idx + 1);
1195 }
1196 return PrototypeDescriptors;
1197}
1198
1199raw_ostream &operator<<(raw_ostream &OS, const RVVIntrinsicRecord &Record) {
1200 OS << "{";
1201 OS << "\"" << Record.Name << "\",";
1202 if (Record.OverloadedName == nullptr ||
1203 StringRef(Record.OverloadedName).empty())
1204 OS << "nullptr,";
1205 else
1206 OS << "\"" << Record.OverloadedName << "\",";
1207 OS << Record.PrototypeIndex << ",";
1208 OS << Record.SuffixIndex << ",";
1209 OS << Record.OverloadedSuffixIndex << ",";
1210 OS << (int)Record.PrototypeLength << ",";
1211 OS << (int)Record.SuffixLength << ",";
1212 OS << (int)Record.OverloadedSuffixSize << ",";
1213 OS << Record.RequiredExtensions << ",";
1214 OS << (int)Record.TypeRangeMask << ",";
1215 OS << (int)Record.Log2LMULMask << ",";
1216 OS << (int)Record.NF << ",";
1217 OS << (int)Record.HasMasked << ",";
1218 OS << (int)Record.HasVL << ",";
1219 OS << (int)Record.HasMaskedOffOperand << ",";
1220 OS << (int)Record.HasTailPolicy << ",";
1221 OS << (int)Record.HasMaskPolicy << ",";
1222 OS << (int)Record.HasFRMRoundModeOp << ",";
1223 OS << (int)Record.IsTuple << ",";
1224 OS << (int)Record.UnMaskedPolicyScheme << ",";
1225 OS << (int)Record.MaskedPolicyScheme << ",";
1226 OS << "},\n";
1227 return OS;
1228}
1229
1230} // end namespace RISCV
1231} // end namespace clang
#define V(N, I)
Definition: ASTContext.h:3341
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:31
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:9098
__device__ int
Complex values, per C99 6.2.5p11.
Definition: Type.h:3134
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, 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:1829
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.
const FunctionProtoType * T
Diagnostic wrappers for TextAPI types for error reporting.
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