35#include "llvm/IR/IntrinsicsAArch64.h"
36#include "llvm/IR/IntrinsicsAMDGPU.h"
37#include "llvm/IR/IntrinsicsARM.h"
38#include "llvm/IR/IntrinsicsNVPTX.h"
39#include "llvm/IR/IntrinsicsRISCV.h"
40#include "llvm/IR/IntrinsicsWebAssembly.h"
41#include "llvm/IR/IntrinsicsX86.h"
64 cl::desc(
"Disable autoupgrade of debug info"));
83 Type *Arg0Type =
F->getFunctionType()->getParamType(0);
98 Type *LastArgType =
F->getFunctionType()->getParamType(
99 F->getFunctionType()->getNumParams() - 1);
114 if (
F->getReturnType()->isVectorTy())
127 Type *Arg1Type =
F->getFunctionType()->getParamType(1);
128 Type *Arg2Type =
F->getFunctionType()->getParamType(2);
145 Type *Arg1Type =
F->getFunctionType()->getParamType(1);
146 Type *Arg2Type =
F->getFunctionType()->getParamType(2);
160 if (
F->getReturnType()->getScalarType()->isBFloatTy())
170 if (
F->getFunctionType()->getParamType(1)->getScalarType()->isBFloatTy())
184 if (Name.consume_front(
"avx."))
185 return (Name.starts_with(
"blend.p") ||
186 Name ==
"cvt.ps2.pd.256" ||
187 Name ==
"cvtdq2.pd.256" ||
188 Name ==
"cvtdq2.ps.256" ||
189 Name.starts_with(
"movnt.") ||
190 Name.starts_with(
"sqrt.p") ||
191 Name.starts_with(
"storeu.") ||
192 Name.starts_with(
"vbroadcast.s") ||
193 Name.starts_with(
"vbroadcastf128") ||
194 Name.starts_with(
"vextractf128.") ||
195 Name.starts_with(
"vinsertf128.") ||
196 Name.starts_with(
"vperm2f128.") ||
197 Name.starts_with(
"vpermil."));
199 if (Name.consume_front(
"avx2."))
200 return (Name ==
"movntdqa" ||
201 Name.starts_with(
"pabs.") ||
202 Name.starts_with(
"padds.") ||
203 Name.starts_with(
"paddus.") ||
204 Name.starts_with(
"pblendd.") ||
206 Name.starts_with(
"pbroadcast") ||
207 Name.starts_with(
"pcmpeq.") ||
208 Name.starts_with(
"pcmpgt.") ||
209 Name.starts_with(
"pmax") ||
210 Name.starts_with(
"pmin") ||
211 Name.starts_with(
"pmovsx") ||
212 Name.starts_with(
"pmovzx") ||
214 Name ==
"pmulu.dq" ||
215 Name.starts_with(
"psll.dq") ||
216 Name.starts_with(
"psrl.dq") ||
217 Name.starts_with(
"psubs.") ||
218 Name.starts_with(
"psubus.") ||
219 Name.starts_with(
"vbroadcast") ||
220 Name ==
"vbroadcasti128" ||
221 Name ==
"vextracti128" ||
222 Name ==
"vinserti128" ||
223 Name ==
"vperm2i128");
225 if (Name.consume_front(
"avx512.")) {
226 if (Name.consume_front(
"mask."))
228 return (Name.starts_with(
"add.p") ||
229 Name.starts_with(
"and.") ||
230 Name.starts_with(
"andn.") ||
231 Name.starts_with(
"broadcast.s") ||
232 Name.starts_with(
"broadcastf32x4.") ||
233 Name.starts_with(
"broadcastf32x8.") ||
234 Name.starts_with(
"broadcastf64x2.") ||
235 Name.starts_with(
"broadcastf64x4.") ||
236 Name.starts_with(
"broadcasti32x4.") ||
237 Name.starts_with(
"broadcasti32x8.") ||
238 Name.starts_with(
"broadcasti64x2.") ||
239 Name.starts_with(
"broadcasti64x4.") ||
240 Name.starts_with(
"cmp.b") ||
241 Name.starts_with(
"cmp.d") ||
242 Name.starts_with(
"cmp.q") ||
243 Name.starts_with(
"cmp.w") ||
244 Name.starts_with(
"compress.b") ||
245 Name.starts_with(
"compress.d") ||
246 Name.starts_with(
"compress.p") ||
247 Name.starts_with(
"compress.q") ||
248 Name.starts_with(
"compress.store.") ||
249 Name.starts_with(
"compress.w") ||
250 Name.starts_with(
"conflict.") ||
251 Name.starts_with(
"cvtdq2pd.") ||
252 Name.starts_with(
"cvtdq2ps.") ||
253 Name ==
"cvtpd2dq.256" ||
254 Name ==
"cvtpd2ps.256" ||
255 Name ==
"cvtps2pd.128" ||
256 Name ==
"cvtps2pd.256" ||
257 Name.starts_with(
"cvtqq2pd.") ||
258 Name ==
"cvtqq2ps.256" ||
259 Name ==
"cvtqq2ps.512" ||
260 Name ==
"cvttpd2dq.256" ||
261 Name ==
"cvttps2dq.128" ||
262 Name ==
"cvttps2dq.256" ||
263 Name.starts_with(
"cvtudq2pd.") ||
264 Name.starts_with(
"cvtudq2ps.") ||
265 Name.starts_with(
"cvtuqq2pd.") ||
266 Name ==
"cvtuqq2ps.256" ||
267 Name ==
"cvtuqq2ps.512" ||
268 Name.starts_with(
"dbpsadbw.") ||
269 Name.starts_with(
"div.p") ||
270 Name.starts_with(
"expand.b") ||
271 Name.starts_with(
"expand.d") ||
272 Name.starts_with(
"expand.load.") ||
273 Name.starts_with(
"expand.p") ||
274 Name.starts_with(
"expand.q") ||
275 Name.starts_with(
"expand.w") ||
276 Name.starts_with(
"fpclass.p") ||
277 Name.starts_with(
"insert") ||
278 Name.starts_with(
"load.") ||
279 Name.starts_with(
"loadu.") ||
280 Name.starts_with(
"lzcnt.") ||
281 Name.starts_with(
"max.p") ||
282 Name.starts_with(
"min.p") ||
283 Name.starts_with(
"movddup") ||
284 Name.starts_with(
"move.s") ||
285 Name.starts_with(
"movshdup") ||
286 Name.starts_with(
"movsldup") ||
287 Name.starts_with(
"mul.p") ||
288 Name.starts_with(
"or.") ||
289 Name.starts_with(
"pabs.") ||
290 Name.starts_with(
"packssdw.") ||
291 Name.starts_with(
"packsswb.") ||
292 Name.starts_with(
"packusdw.") ||
293 Name.starts_with(
"packuswb.") ||
294 Name.starts_with(
"padd.") ||
295 Name.starts_with(
"padds.") ||
296 Name.starts_with(
"paddus.") ||
297 Name.starts_with(
"palignr.") ||
298 Name.starts_with(
"pand.") ||
299 Name.starts_with(
"pandn.") ||
300 Name.starts_with(
"pavg") ||
301 Name.starts_with(
"pbroadcast") ||
302 Name.starts_with(
"pcmpeq.") ||
303 Name.starts_with(
"pcmpgt.") ||
304 Name.starts_with(
"perm.df.") ||
305 Name.starts_with(
"perm.di.") ||
306 Name.starts_with(
"permvar.") ||
307 Name.starts_with(
"pmaddubs.w.") ||
308 Name.starts_with(
"pmaddw.d.") ||
309 Name.starts_with(
"pmax") ||
310 Name.starts_with(
"pmin") ||
311 Name ==
"pmov.qd.256" ||
312 Name ==
"pmov.qd.512" ||
313 Name ==
"pmov.wb.256" ||
314 Name ==
"pmov.wb.512" ||
315 Name.starts_with(
"pmovsx") ||
316 Name.starts_with(
"pmovzx") ||
317 Name.starts_with(
"pmul.dq.") ||
318 Name.starts_with(
"pmul.hr.sw.") ||
319 Name.starts_with(
"pmulh.w.") ||
320 Name.starts_with(
"pmulhu.w.") ||
321 Name.starts_with(
"pmull.") ||
322 Name.starts_with(
"pmultishift.qb.") ||
323 Name.starts_with(
"pmulu.dq.") ||
324 Name.starts_with(
"por.") ||
325 Name.starts_with(
"prol.") ||
326 Name.starts_with(
"prolv.") ||
327 Name.starts_with(
"pror.") ||
328 Name.starts_with(
"prorv.") ||
329 Name.starts_with(
"pshuf.b.") ||
330 Name.starts_with(
"pshuf.d.") ||
331 Name.starts_with(
"pshufh.w.") ||
332 Name.starts_with(
"pshufl.w.") ||
333 Name.starts_with(
"psll.d") ||
334 Name.starts_with(
"psll.q") ||
335 Name.starts_with(
"psll.w") ||
336 Name.starts_with(
"pslli") ||
337 Name.starts_with(
"psllv") ||
338 Name.starts_with(
"psra.d") ||
339 Name.starts_with(
"psra.q") ||
340 Name.starts_with(
"psra.w") ||
341 Name.starts_with(
"psrai") ||
342 Name.starts_with(
"psrav") ||
343 Name.starts_with(
"psrl.d") ||
344 Name.starts_with(
"psrl.q") ||
345 Name.starts_with(
"psrl.w") ||
346 Name.starts_with(
"psrli") ||
347 Name.starts_with(
"psrlv") ||
348 Name.starts_with(
"psub.") ||
349 Name.starts_with(
"psubs.") ||
350 Name.starts_with(
"psubus.") ||
351 Name.starts_with(
"pternlog.") ||
352 Name.starts_with(
"punpckh") ||
353 Name.starts_with(
"punpckl") ||
354 Name.starts_with(
"pxor.") ||
355 Name.starts_with(
"shuf.f") ||
356 Name.starts_with(
"shuf.i") ||
357 Name.starts_with(
"shuf.p") ||
358 Name.starts_with(
"sqrt.p") ||
359 Name.starts_with(
"store.b.") ||
360 Name.starts_with(
"store.d.") ||
361 Name.starts_with(
"store.p") ||
362 Name.starts_with(
"store.q.") ||
363 Name.starts_with(
"store.w.") ||
364 Name ==
"store.ss" ||
365 Name.starts_with(
"storeu.") ||
366 Name.starts_with(
"sub.p") ||
367 Name.starts_with(
"ucmp.") ||
368 Name.starts_with(
"unpckh.") ||
369 Name.starts_with(
"unpckl.") ||
370 Name.starts_with(
"valign.") ||
371 Name ==
"vcvtph2ps.128" ||
372 Name ==
"vcvtph2ps.256" ||
373 Name.starts_with(
"vextract") ||
374 Name.starts_with(
"vfmadd.") ||
375 Name.starts_with(
"vfmaddsub.") ||
376 Name.starts_with(
"vfnmadd.") ||
377 Name.starts_with(
"vfnmsub.") ||
378 Name.starts_with(
"vpdpbusd.") ||
379 Name.starts_with(
"vpdpbusds.") ||
380 Name.starts_with(
"vpdpwssd.") ||
381 Name.starts_with(
"vpdpwssds.") ||
382 Name.starts_with(
"vpermi2var.") ||
383 Name.starts_with(
"vpermil.p") ||
384 Name.starts_with(
"vpermilvar.") ||
385 Name.starts_with(
"vpermt2var.") ||
386 Name.starts_with(
"vpmadd52") ||
387 Name.starts_with(
"vpshld.") ||
388 Name.starts_with(
"vpshldv.") ||
389 Name.starts_with(
"vpshrd.") ||
390 Name.starts_with(
"vpshrdv.") ||
391 Name.starts_with(
"vpshufbitqmb.") ||
392 Name.starts_with(
"xor."));
394 if (Name.consume_front(
"mask3."))
396 return (Name.starts_with(
"vfmadd.") ||
397 Name.starts_with(
"vfmaddsub.") ||
398 Name.starts_with(
"vfmsub.") ||
399 Name.starts_with(
"vfmsubadd.") ||
400 Name.starts_with(
"vfnmsub."));
402 if (Name.consume_front(
"maskz."))
404 return (Name.starts_with(
"pternlog.") ||
405 Name.starts_with(
"vfmadd.") ||
406 Name.starts_with(
"vfmaddsub.") ||
407 Name.starts_with(
"vpdpbusd.") ||
408 Name.starts_with(
"vpdpbusds.") ||
409 Name.starts_with(
"vpdpwssd.") ||
410 Name.starts_with(
"vpdpwssds.") ||
411 Name.starts_with(
"vpermt2var.") ||
412 Name.starts_with(
"vpmadd52") ||
413 Name.starts_with(
"vpshldv.") ||
414 Name.starts_with(
"vpshrdv."));
417 return (Name ==
"movntdqa" ||
418 Name ==
"pmul.dq.512" ||
419 Name ==
"pmulu.dq.512" ||
420 Name.starts_with(
"broadcastm") ||
421 Name.starts_with(
"cmp.p") ||
422 Name.starts_with(
"cvtb2mask.") ||
423 Name.starts_with(
"cvtd2mask.") ||
424 Name.starts_with(
"cvtmask2") ||
425 Name.starts_with(
"cvtq2mask.") ||
426 Name ==
"cvtusi2sd" ||
427 Name.starts_with(
"cvtw2mask.") ||
432 Name ==
"kortestc.w" ||
433 Name ==
"kortestz.w" ||
434 Name.starts_with(
"kunpck") ||
437 Name.starts_with(
"padds.") ||
438 Name.starts_with(
"pbroadcast") ||
439 Name.starts_with(
"prol") ||
440 Name.starts_with(
"pror") ||
441 Name.starts_with(
"psll.dq") ||
442 Name.starts_with(
"psrl.dq") ||
443 Name.starts_with(
"psubs.") ||
444 Name.starts_with(
"ptestm") ||
445 Name.starts_with(
"ptestnm") ||
446 Name.starts_with(
"storent.") ||
447 Name.starts_with(
"vbroadcast.s") ||
448 Name.starts_with(
"vpshld.") ||
449 Name.starts_with(
"vpshrd."));
452 if (Name.consume_front(
"fma."))
453 return (Name.starts_with(
"vfmadd.") ||
454 Name.starts_with(
"vfmsub.") ||
455 Name.starts_with(
"vfmsubadd.") ||
456 Name.starts_with(
"vfnmadd.") ||
457 Name.starts_with(
"vfnmsub."));
459 if (Name.consume_front(
"fma4."))
460 return Name.starts_with(
"vfmadd.s");
462 if (Name.consume_front(
"sse."))
463 return (Name ==
"add.ss" ||
464 Name ==
"cvtsi2ss" ||
465 Name ==
"cvtsi642ss" ||
468 Name.starts_with(
"sqrt.p") ||
470 Name.starts_with(
"storeu.") ||
473 if (Name.consume_front(
"sse2."))
474 return (Name ==
"add.sd" ||
475 Name ==
"cvtdq2pd" ||
476 Name ==
"cvtdq2ps" ||
477 Name ==
"cvtps2pd" ||
478 Name ==
"cvtsi2sd" ||
479 Name ==
"cvtsi642sd" ||
480 Name ==
"cvtss2sd" ||
483 Name.starts_with(
"padds.") ||
484 Name.starts_with(
"paddus.") ||
485 Name.starts_with(
"pcmpeq.") ||
486 Name.starts_with(
"pcmpgt.") ||
491 Name ==
"pmulu.dq" ||
492 Name.starts_with(
"pshuf") ||
493 Name.starts_with(
"psll.dq") ||
494 Name.starts_with(
"psrl.dq") ||
495 Name.starts_with(
"psubs.") ||
496 Name.starts_with(
"psubus.") ||
497 Name.starts_with(
"sqrt.p") ||
499 Name ==
"storel.dq" ||
500 Name.starts_with(
"storeu.") ||
503 if (Name.consume_front(
"sse41."))
504 return (Name.starts_with(
"blendp") ||
505 Name ==
"movntdqa" ||
515 Name.starts_with(
"pmovsx") ||
516 Name.starts_with(
"pmovzx") ||
519 if (Name.consume_front(
"sse42."))
520 return Name ==
"crc32.64.8";
522 if (Name.consume_front(
"sse4a."))
523 return Name.starts_with(
"movnt.");
525 if (Name.consume_front(
"ssse3."))
526 return (Name ==
"pabs.b.128" ||
527 Name ==
"pabs.d.128" ||
528 Name ==
"pabs.w.128");
530 if (Name.consume_front(
"xop."))
531 return (Name ==
"vpcmov" ||
532 Name ==
"vpcmov.256" ||
533 Name.starts_with(
"vpcom") ||
534 Name.starts_with(
"vprot"));
536 if (Name.consume_front(
"bmi."))
537 return (Name.starts_with(
"pdep.") ||
538 Name.starts_with(
"pext."));
540 return (Name ==
"addcarry.u32" ||
541 Name ==
"addcarry.u64" ||
542 Name ==
"addcarryx.u32" ||
543 Name ==
"addcarryx.u64" ||
544 Name ==
"subborrow.u32" ||
545 Name ==
"subborrow.u64" ||
546 Name.starts_with(
"vcvtph2ps."));
552 if (!Name.consume_front(
"x86."))
560 if (Name ==
"rdtscp") {
562 if (
F->getFunctionType()->getNumParams() == 0)
567 Intrinsic::x86_rdtscp);
574 if (Name.consume_front(
"sse41.ptest")) {
576 .
Case(
"c", Intrinsic::x86_sse41_ptestc)
577 .
Case(
"z", Intrinsic::x86_sse41_ptestz)
578 .
Case(
"nzc", Intrinsic::x86_sse41_ptestnzc)
591 .
Case(
"sse41.insertps", Intrinsic::x86_sse41_insertps)
592 .
Case(
"sse41.dppd", Intrinsic::x86_sse41_dppd)
593 .
Case(
"sse41.dpps", Intrinsic::x86_sse41_dpps)
594 .
Case(
"sse41.mpsadbw", Intrinsic::x86_sse41_mpsadbw)
595 .
Case(
"avx.dp.ps.256", Intrinsic::x86_avx_dp_ps_256)
596 .
Case(
"avx2.mpsadbw", Intrinsic::x86_avx2_mpsadbw)
601 if (Name.consume_front(
"avx512.")) {
602 if (Name.consume_front(
"mask.cmp.")) {
605 .
Case(
"pd.128", Intrinsic::x86_avx512_mask_cmp_pd_128)
606 .
Case(
"pd.256", Intrinsic::x86_avx512_mask_cmp_pd_256)
607 .
Case(
"pd.512", Intrinsic::x86_avx512_mask_cmp_pd_512)
608 .
Case(
"ps.128", Intrinsic::x86_avx512_mask_cmp_ps_128)
609 .
Case(
"ps.256", Intrinsic::x86_avx512_mask_cmp_ps_256)
610 .
Case(
"ps.512", Intrinsic::x86_avx512_mask_cmp_ps_512)
614 }
else if (Name.starts_with(
"vpdpbusd.") ||
615 Name.starts_with(
"vpdpbusds.")) {
618 .
Case(
"vpdpbusd.128", Intrinsic::x86_avx512_vpdpbusd_128)
619 .
Case(
"vpdpbusd.256", Intrinsic::x86_avx512_vpdpbusd_256)
620 .
Case(
"vpdpbusd.512", Intrinsic::x86_avx512_vpdpbusd_512)
621 .
Case(
"vpdpbusds.128", Intrinsic::x86_avx512_vpdpbusds_128)
622 .
Case(
"vpdpbusds.256", Intrinsic::x86_avx512_vpdpbusds_256)
623 .
Case(
"vpdpbusds.512", Intrinsic::x86_avx512_vpdpbusds_512)
627 }
else if (Name.starts_with(
"vpdpwssd.") ||
628 Name.starts_with(
"vpdpwssds.")) {
631 .
Case(
"vpdpwssd.128", Intrinsic::x86_avx512_vpdpwssd_128)
632 .
Case(
"vpdpwssd.256", Intrinsic::x86_avx512_vpdpwssd_256)
633 .
Case(
"vpdpwssd.512", Intrinsic::x86_avx512_vpdpwssd_512)
634 .
Case(
"vpdpwssds.128", Intrinsic::x86_avx512_vpdpwssds_128)
635 .
Case(
"vpdpwssds.256", Intrinsic::x86_avx512_vpdpwssds_256)
636 .
Case(
"vpdpwssds.512", Intrinsic::x86_avx512_vpdpwssds_512)
644 if (Name.consume_front(
"avx2.")) {
645 if (Name.consume_front(
"vpdpb")) {
648 .
Case(
"ssd.128", Intrinsic::x86_avx2_vpdpbssd_128)
649 .
Case(
"ssd.256", Intrinsic::x86_avx2_vpdpbssd_256)
650 .
Case(
"ssds.128", Intrinsic::x86_avx2_vpdpbssds_128)
651 .
Case(
"ssds.256", Intrinsic::x86_avx2_vpdpbssds_256)
652 .
Case(
"sud.128", Intrinsic::x86_avx2_vpdpbsud_128)
653 .
Case(
"sud.256", Intrinsic::x86_avx2_vpdpbsud_256)
654 .
Case(
"suds.128", Intrinsic::x86_avx2_vpdpbsuds_128)
655 .
Case(
"suds.256", Intrinsic::x86_avx2_vpdpbsuds_256)
656 .
Case(
"uud.128", Intrinsic::x86_avx2_vpdpbuud_128)
657 .
Case(
"uud.256", Intrinsic::x86_avx2_vpdpbuud_256)
658 .
Case(
"uuds.128", Intrinsic::x86_avx2_vpdpbuuds_128)
659 .
Case(
"uuds.256", Intrinsic::x86_avx2_vpdpbuuds_256)
663 }
else if (Name.consume_front(
"vpdpw")) {
666 .
Case(
"sud.128", Intrinsic::x86_avx2_vpdpwsud_128)
667 .
Case(
"sud.256", Intrinsic::x86_avx2_vpdpwsud_256)
668 .
Case(
"suds.128", Intrinsic::x86_avx2_vpdpwsuds_128)
669 .
Case(
"suds.256", Intrinsic::x86_avx2_vpdpwsuds_256)
670 .
Case(
"usd.128", Intrinsic::x86_avx2_vpdpwusd_128)
671 .
Case(
"usd.256", Intrinsic::x86_avx2_vpdpwusd_256)
672 .
Case(
"usds.128", Intrinsic::x86_avx2_vpdpwusds_128)
673 .
Case(
"usds.256", Intrinsic::x86_avx2_vpdpwusds_256)
674 .
Case(
"uud.128", Intrinsic::x86_avx2_vpdpwuud_128)
675 .
Case(
"uud.256", Intrinsic::x86_avx2_vpdpwuud_256)
676 .
Case(
"uuds.128", Intrinsic::x86_avx2_vpdpwuuds_128)
677 .
Case(
"uuds.256", Intrinsic::x86_avx2_vpdpwuuds_256)
685 if (Name.consume_front(
"avx10.")) {
686 if (Name.consume_front(
"vpdpb")) {
689 .
Case(
"ssd.512", Intrinsic::x86_avx10_vpdpbssd_512)
690 .
Case(
"ssds.512", Intrinsic::x86_avx10_vpdpbssds_512)
691 .
Case(
"sud.512", Intrinsic::x86_avx10_vpdpbsud_512)
692 .
Case(
"suds.512", Intrinsic::x86_avx10_vpdpbsuds_512)
693 .
Case(
"uud.512", Intrinsic::x86_avx10_vpdpbuud_512)
694 .
Case(
"uuds.512", Intrinsic::x86_avx10_vpdpbuuds_512)
698 }
else if (Name.consume_front(
"vpdpw")) {
700 .
Case(
"sud.512", Intrinsic::x86_avx10_vpdpwsud_512)
701 .
Case(
"suds.512", Intrinsic::x86_avx10_vpdpwsuds_512)
702 .
Case(
"usd.512", Intrinsic::x86_avx10_vpdpwusd_512)
703 .
Case(
"usds.512", Intrinsic::x86_avx10_vpdpwusds_512)
704 .
Case(
"uud.512", Intrinsic::x86_avx10_vpdpwuud_512)
705 .
Case(
"uuds.512", Intrinsic::x86_avx10_vpdpwuuds_512)
713 if (Name.consume_front(
"avx512bf16.")) {
716 .
Case(
"cvtne2ps2bf16.128",
717 Intrinsic::x86_avx512bf16_cvtne2ps2bf16_128)
718 .
Case(
"cvtne2ps2bf16.256",
719 Intrinsic::x86_avx512bf16_cvtne2ps2bf16_256)
720 .
Case(
"cvtne2ps2bf16.512",
721 Intrinsic::x86_avx512bf16_cvtne2ps2bf16_512)
722 .
Case(
"mask.cvtneps2bf16.128",
723 Intrinsic::x86_avx512bf16_mask_cvtneps2bf16_128)
724 .
Case(
"cvtneps2bf16.256",
725 Intrinsic::x86_avx512bf16_cvtneps2bf16_256)
726 .
Case(
"cvtneps2bf16.512",
727 Intrinsic::x86_avx512bf16_cvtneps2bf16_512)
734 .
Case(
"dpbf16ps.128", Intrinsic::x86_avx512bf16_dpbf16ps_128)
735 .
Case(
"dpbf16ps.256", Intrinsic::x86_avx512bf16_dpbf16ps_256)
736 .
Case(
"dpbf16ps.512", Intrinsic::x86_avx512bf16_dpbf16ps_512)
743 if (Name.consume_front(
"xop.")) {
745 if (Name.starts_with(
"vpermil2")) {
748 auto Idx =
F->getFunctionType()->getParamType(2);
749 if (Idx->isFPOrFPVectorTy()) {
750 unsigned IdxSize = Idx->getPrimitiveSizeInBits();
751 unsigned EltSize = Idx->getScalarSizeInBits();
752 if (EltSize == 64 && IdxSize == 128)
753 ID = Intrinsic::x86_xop_vpermil2pd;
754 else if (EltSize == 32 && IdxSize == 128)
755 ID = Intrinsic::x86_xop_vpermil2ps;
756 else if (EltSize == 64 && IdxSize == 256)
757 ID = Intrinsic::x86_xop_vpermil2pd_256;
759 ID = Intrinsic::x86_xop_vpermil2ps_256;
761 }
else if (
F->arg_size() == 2)
764 .
Case(
"vfrcz.ss", Intrinsic::x86_xop_vfrcz_ss)
765 .
Case(
"vfrcz.sd", Intrinsic::x86_xop_vfrcz_sd)
776 if (Name ==
"seh.recoverfp") {
778 Intrinsic::eh_recoverfp);
790 if (Name.starts_with(
"rbit")) {
793 F->getParent(), Intrinsic::bitreverse,
F->arg_begin()->getType());
797 if (Name ==
"thread.pointer") {
800 F->getParent(), Intrinsic::thread_pointer,
F->getReturnType());
804 bool Neon = Name.consume_front(
"neon.");
809 if (Name.consume_front(
"bfdot.")) {
813 .
Cases({
"v2f32.v8i8",
"v4f32.v16i8"},
818 size_t OperandWidth =
F->getReturnType()->getPrimitiveSizeInBits();
819 assert((OperandWidth == 64 || OperandWidth == 128) &&
820 "Unexpected operand width");
822 std::array<Type *, 2> Tys{
833 if (Name.consume_front(
"bfm")) {
835 if (Name.consume_back(
".v4f32.v16i8")) {
881 F->arg_begin()->getType());
885 if (Name.consume_front(
"vst")) {
887 static const Regex vstRegex(
"^([1234]|[234]lane)\\.v[a-z0-9]*$");
891 Intrinsic::arm_neon_vst1, Intrinsic::arm_neon_vst2,
892 Intrinsic::arm_neon_vst3, Intrinsic::arm_neon_vst4};
895 Intrinsic::arm_neon_vst2lane, Intrinsic::arm_neon_vst3lane,
896 Intrinsic::arm_neon_vst4lane};
898 auto fArgs =
F->getFunctionType()->params();
899 Type *Tys[] = {fArgs[0], fArgs[1]};
902 F->getParent(), StoreInts[fArgs.size() - 3], Tys);
905 F->getParent(), StoreLaneInts[fArgs.size() - 5], Tys);
914 if (Name.consume_front(
"mve.")) {
916 if (Name ==
"vctp64") {
926 if (Name.starts_with(
"vrintn.v")) {
928 F->getParent(), Intrinsic::roundeven,
F->arg_begin()->getType());
933 if (Name.consume_back(
".v4i1")) {
935 if (Name.consume_back(
".predicated.v2i64.v4i32"))
937 return Name ==
"mull.int" || Name ==
"vqdmull";
939 if (Name.consume_back(
".v2i64")) {
941 bool IsGather = Name.consume_front(
"vldr.gather.");
942 if (IsGather || Name.consume_front(
"vstr.scatter.")) {
943 if (Name.consume_front(
"base.")) {
945 Name.consume_front(
"wb.");
948 return Name ==
"predicated.v2i64";
951 if (Name.consume_front(
"offset.predicated."))
952 return Name == (IsGather ?
"v2i64.p0i64" :
"p0i64.v2i64") ||
953 Name == (IsGather ?
"v2i64.p0" :
"p0.v2i64");
966 if (Name.consume_front(
"cde.vcx")) {
968 if (Name.consume_back(
".predicated.v2i64.v4i1"))
970 return Name ==
"1q" || Name ==
"1qa" || Name ==
"2q" || Name ==
"2qa" ||
971 Name ==
"3q" || Name ==
"3qa";
985 F->arg_begin()->getType());
989 if (Name.starts_with(
"addp")) {
991 if (
F->arg_size() != 2)
994 if (Ty && Ty->getElementType()->isFloatingPointTy()) {
996 F->getParent(), Intrinsic::aarch64_neon_faddp, Ty);
1002 if (Name.starts_with(
"bfcvt")) {
1008 if (Name ==
"vcvtfp2hf" || Name ==
"vcvthf2fp") {
1015 if (Name.consume_front(
"sve.")) {
1017 if (Name.consume_front(
"bf")) {
1018 if (Name ==
"mmla") {
1019 Type *Tys[] = {
F->getReturnType(),
1020 std::next(
F->arg_begin())->getType()};
1022 F->getParent(), Intrinsic::aarch64_sve_fmmla, Tys);
1025 if (Name.consume_back(
".lane")) {
1029 .
Case(
"dot", Intrinsic::aarch64_sve_bfdot_lane_v2)
1030 .
Case(
"mlalb", Intrinsic::aarch64_sve_bfmlalb_lane_v2)
1031 .
Case(
"mlalt", Intrinsic::aarch64_sve_bfmlalt_lane_v2)
1043 if (Name ==
"fcvt.bf16f32" || Name ==
"fcvtnt.bf16f32") {
1048 if (Name.consume_front(
"addqv")) {
1050 if (!
F->getReturnType()->isFPOrFPVectorTy())
1053 auto Args =
F->getFunctionType()->params();
1054 Type *Tys[] = {
F->getReturnType(), Args[1]};
1056 F->getParent(), Intrinsic::aarch64_sve_faddqv, Tys);
1060 if (Name.consume_front(
"ld")) {
1062 static const Regex LdRegex(
"^[234](.nxv[a-z0-9]+|$)");
1063 if (LdRegex.
match(Name)) {
1069 "Expected 2 arguments for ld* intrinsic.");
1070 Type *PtrTy =
F->getArg(1)->getType();
1073 Intrinsic::aarch64_sve_ld2_sret,
1074 Intrinsic::aarch64_sve_ld3_sret,
1075 Intrinsic::aarch64_sve_ld4_sret,
1078 F->getParent(), LoadIDs[Name[0] -
'2'], {Ty, PtrTy});
1084 if (Name.consume_front(
"tuple.")) {
1086 if (Name.starts_with(
"get")) {
1088 Type *Tys[] = {
F->getReturnType(),
F->arg_begin()->getType()};
1090 F->getParent(), Intrinsic::vector_extract, Tys);
1094 if (Name.starts_with(
"set")) {
1096 auto Args =
F->getFunctionType()->params();
1097 Type *Tys[] = {Args[0], Args[2], Args[1]};
1099 F->getParent(), Intrinsic::vector_insert, Tys);
1103 static const Regex CreateTupleRegex(
"^create[234](.nxv[a-z0-9]+|$)");
1104 if (CreateTupleRegex.
match(Name)) {
1106 auto Args =
F->getFunctionType()->params();
1107 Type *Tys[] = {
F->getReturnType(), Args[1]};
1109 F->getParent(), Intrinsic::vector_insert, Tys);
1115 if (Name.starts_with(
"rev.nxv")) {
1118 F->getParent(), Intrinsic::vector_reverse,
F->getReturnType());
1124 if (Name.consume_front(
"sme.")) {
1126 if (Name.consume_front(
"ftmopa.")) {
1131 .
Case(
"za16.nxv16i8", Intrinsic::aarch64_sme_fp8_ftmopa_za16)
1132 .
Case(
"za32.nxv16i8", Intrinsic::aarch64_sme_fp8_ftmopa_za32)
1149 if (Name.consume_front(
"cp.async.bulk.tensor.g2s.")) {
1153 Intrinsic::nvvm_cp_async_bulk_tensor_g2s_im2col_3d)
1155 Intrinsic::nvvm_cp_async_bulk_tensor_g2s_im2col_4d)
1157 Intrinsic::nvvm_cp_async_bulk_tensor_g2s_im2col_5d)
1158 .
Case(
"tile.1d", Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_1d)
1159 .
Case(
"tile.2d", Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_2d)
1160 .
Case(
"tile.3d", Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_3d)
1161 .
Case(
"tile.4d", Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_4d)
1162 .
Case(
"tile.5d", Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_5d)
1171 if (
F->getArg(0)->getType()->getPointerAddressSpace() ==
1185 size_t FlagStartIndex =
F->getFunctionType()->getNumParams() - 3;
1186 Type *ArgType =
F->getFunctionType()->getParamType(FlagStartIndex);
1196 if (Name.consume_front(
"mapa.shared.cluster"))
1197 if (
F->getReturnType()->getPointerAddressSpace() ==
1199 return Intrinsic::nvvm_mapa_shared_cluster;
1201 if (Name.consume_front(
"cp.async.bulk.")) {
1204 .
Case(
"global.to.shared.cluster",
1205 Intrinsic::nvvm_cp_async_bulk_global_to_shared_cluster)
1206 .
Case(
"shared.cta.to.cluster",
1207 Intrinsic::nvvm_cp_async_bulk_shared_cta_to_cluster)
1211 if (
F->getArg(0)->getType()->getPointerAddressSpace() ==
1220 if (Name.consume_front(
"fma.rn."))
1222 .
Case(
"bf16", Intrinsic::nvvm_fma_rn_bf16)
1223 .
Case(
"bf16x2", Intrinsic::nvvm_fma_rn_bf16x2)
1224 .
Case(
"relu.bf16", Intrinsic::nvvm_fma_rn_relu_bf16)
1225 .
Case(
"relu.bf16x2", Intrinsic::nvvm_fma_rn_relu_bf16x2)
1228 if (Name.consume_front(
"fmax."))
1230 .
Case(
"bf16", Intrinsic::nvvm_fmax_bf16)
1231 .
Case(
"bf16x2", Intrinsic::nvvm_fmax_bf16x2)
1232 .
Case(
"ftz.bf16", Intrinsic::nvvm_fmax_ftz_bf16)
1233 .
Case(
"ftz.bf16x2", Intrinsic::nvvm_fmax_ftz_bf16x2)
1234 .
Case(
"ftz.nan.bf16", Intrinsic::nvvm_fmax_ftz_nan_bf16)
1235 .
Case(
"ftz.nan.bf16x2", Intrinsic::nvvm_fmax_ftz_nan_bf16x2)
1236 .
Case(
"ftz.nan.xorsign.abs.bf16",
1237 Intrinsic::nvvm_fmax_ftz_nan_xorsign_abs_bf16)
1238 .
Case(
"ftz.nan.xorsign.abs.bf16x2",
1239 Intrinsic::nvvm_fmax_ftz_nan_xorsign_abs_bf16x2)
1240 .
Case(
"ftz.xorsign.abs.bf16", Intrinsic::nvvm_fmax_ftz_xorsign_abs_bf16)
1241 .
Case(
"ftz.xorsign.abs.bf16x2",
1242 Intrinsic::nvvm_fmax_ftz_xorsign_abs_bf16x2)
1243 .
Case(
"nan.bf16", Intrinsic::nvvm_fmax_nan_bf16)
1244 .
Case(
"nan.bf16x2", Intrinsic::nvvm_fmax_nan_bf16x2)
1245 .
Case(
"nan.xorsign.abs.bf16", Intrinsic::nvvm_fmax_nan_xorsign_abs_bf16)
1246 .
Case(
"nan.xorsign.abs.bf16x2",
1247 Intrinsic::nvvm_fmax_nan_xorsign_abs_bf16x2)
1248 .
Case(
"xorsign.abs.bf16", Intrinsic::nvvm_fmax_xorsign_abs_bf16)
1249 .
Case(
"xorsign.abs.bf16x2", Intrinsic::nvvm_fmax_xorsign_abs_bf16x2)
1252 if (Name.consume_front(
"fmin."))
1254 .
Case(
"bf16", Intrinsic::nvvm_fmin_bf16)
1255 .
Case(
"bf16x2", Intrinsic::nvvm_fmin_bf16x2)
1256 .
Case(
"ftz.bf16", Intrinsic::nvvm_fmin_ftz_bf16)
1257 .
Case(
"ftz.bf16x2", Intrinsic::nvvm_fmin_ftz_bf16x2)
1258 .
Case(
"ftz.nan.bf16", Intrinsic::nvvm_fmin_ftz_nan_bf16)
1259 .
Case(
"ftz.nan.bf16x2", Intrinsic::nvvm_fmin_ftz_nan_bf16x2)
1260 .
Case(
"ftz.nan.xorsign.abs.bf16",
1261 Intrinsic::nvvm_fmin_ftz_nan_xorsign_abs_bf16)
1262 .
Case(
"ftz.nan.xorsign.abs.bf16x2",
1263 Intrinsic::nvvm_fmin_ftz_nan_xorsign_abs_bf16x2)
1264 .
Case(
"ftz.xorsign.abs.bf16", Intrinsic::nvvm_fmin_ftz_xorsign_abs_bf16)
1265 .
Case(
"ftz.xorsign.abs.bf16x2",
1266 Intrinsic::nvvm_fmin_ftz_xorsign_abs_bf16x2)
1267 .
Case(
"nan.bf16", Intrinsic::nvvm_fmin_nan_bf16)
1268 .
Case(
"nan.bf16x2", Intrinsic::nvvm_fmin_nan_bf16x2)
1269 .
Case(
"nan.xorsign.abs.bf16", Intrinsic::nvvm_fmin_nan_xorsign_abs_bf16)
1270 .
Case(
"nan.xorsign.abs.bf16x2",
1271 Intrinsic::nvvm_fmin_nan_xorsign_abs_bf16x2)
1272 .
Case(
"xorsign.abs.bf16", Intrinsic::nvvm_fmin_xorsign_abs_bf16)
1273 .
Case(
"xorsign.abs.bf16x2", Intrinsic::nvvm_fmin_xorsign_abs_bf16x2)
1276 if (Name.consume_front(
"neg."))
1278 .
Case(
"bf16", Intrinsic::nvvm_neg_bf16)
1279 .
Case(
"bf16x2", Intrinsic::nvvm_neg_bf16x2)
1286 return Name.consume_front(
"local") || Name.consume_front(
"shared") ||
1287 Name.consume_front(
"global") || Name.consume_front(
"constant") ||
1288 Name.consume_front(
"param");
1294 if (Name.starts_with(
"to.fp16")) {
1298 FuncTy->getReturnType());
1301 if (Name.starts_with(
"from.fp16")) {
1305 FuncTy->getReturnType());
1317 if (Defaults.empty())
1329 if (
F->arg_size() >= FullDecl->
arg_size())
1334 if (
F->arg_size() < FirstDefault)
1342 bool CanUpgradeDebugIntrinsicsToRecords) {
1343 assert(
F &&
"Illegal to upgrade a non-existent Function.");
1348 if (!Name.consume_front(
"llvm.") || Name.empty())
1354 bool IsArm = Name.consume_front(
"arm.");
1355 if (IsArm || Name.consume_front(
"aarch64.")) {
1361 if (Name.consume_front(
"amdgcn.")) {
1362 if (Name ==
"alignbit") {
1365 F->getParent(), Intrinsic::fshr, {F->getReturnType()});
1369 if (Name.consume_front(
"atomic.")) {
1370 if (Name.starts_with(
"inc") || Name.starts_with(
"dec") ||
1371 Name.starts_with(
"cond.sub") || Name.starts_with(
"csub")) {
1380 switch (
F->getIntrinsicID()) {
1384 case Intrinsic::amdgcn_wmma_i32_16x16x64_iu8:
1385 if (
F->arg_size() == 7) {
1390 case Intrinsic::amdgcn_swmmac_i32_16x16x128_iu8:
1391 case Intrinsic::amdgcn_wmma_f32_16x16x4_f32:
1392 case Intrinsic::amdgcn_wmma_f32_16x16x32_bf16:
1393 case Intrinsic::amdgcn_wmma_f32_16x16x32_f16:
1394 case Intrinsic::amdgcn_wmma_f16_16x16x32_f16:
1395 case Intrinsic::amdgcn_wmma_bf16_16x16x32_bf16:
1396 case Intrinsic::amdgcn_wmma_bf16f32_16x16x32_bf16:
1397 if (
F->arg_size() == 8) {
1404 if (Name.consume_front(
"ds.") || Name.consume_front(
"global.atomic.") ||
1405 Name.consume_front(
"flat.atomic.")) {
1406 if (Name.starts_with(
"fadd") ||
1408 (Name.starts_with(
"fmin") && !Name.starts_with(
"fmin.num")) ||
1409 (Name.starts_with(
"fmax") && !Name.starts_with(
"fmax.num"))) {
1417 if (Name.starts_with(
"ldexp.")) {
1420 F->getParent(), Intrinsic::ldexp,
1421 {F->getReturnType(), F->getArg(1)->getType()});
1430 if (
F->arg_size() == 1) {
1431 if (Name.consume_front(
"convert.")) {
1445 F->arg_begin()->getType());
1450 if (
F->arg_size() == 2 && Name ==
"coro.end") {
1453 Intrinsic::coro_end);
1460 if (Name.consume_front(
"dbg.")) {
1462 if (CanUpgradeDebugIntrinsicsToRecords) {
1463 if (Name ==
"addr" || Name ==
"value" || Name ==
"assign" ||
1464 Name ==
"declare" || Name ==
"label") {
1473 if (Name ==
"addr" || (Name ==
"value" &&
F->arg_size() == 4)) {
1476 Intrinsic::dbg_value);
1483 if (Name.consume_front(
"experimental.vector.")) {
1489 .
StartsWith(
"extract.", Intrinsic::vector_extract)
1490 .
StartsWith(
"insert.", Intrinsic::vector_insert)
1491 .
StartsWith(
"reverse.", Intrinsic::vector_reverse)
1492 .
StartsWith(
"interleave2.", Intrinsic::vector_interleave2)
1493 .
StartsWith(
"deinterleave2.", Intrinsic::vector_deinterleave2)
1495 Intrinsic::vector_partial_reduce_add)
1498 const auto *FT =
F->getFunctionType();
1500 if (
ID == Intrinsic::vector_extract ||
1501 ID == Intrinsic::vector_interleave2)
1504 if (
ID != Intrinsic::vector_interleave2)
1506 if (
ID == Intrinsic::vector_insert ||
1507 ID == Intrinsic::vector_partial_reduce_add)
1515 if (Name.consume_front(
"reduce.")) {
1517 static const Regex R(
"^([a-z]+)\\.[a-z][0-9]+");
1518 if (R.match(Name, &
Groups))
1520 .
Case(
"add", Intrinsic::vector_reduce_add)
1521 .
Case(
"mul", Intrinsic::vector_reduce_mul)
1522 .
Case(
"and", Intrinsic::vector_reduce_and)
1523 .
Case(
"or", Intrinsic::vector_reduce_or)
1524 .
Case(
"xor", Intrinsic::vector_reduce_xor)
1525 .
Case(
"smax", Intrinsic::vector_reduce_smax)
1526 .
Case(
"smin", Intrinsic::vector_reduce_smin)
1527 .
Case(
"umax", Intrinsic::vector_reduce_umax)
1528 .
Case(
"umin", Intrinsic::vector_reduce_umin)
1529 .
Case(
"fmax", Intrinsic::vector_reduce_fmax)
1530 .
Case(
"fmin", Intrinsic::vector_reduce_fmin)
1535 static const Regex R2(
"^v2\\.([a-z]+)\\.[fi][0-9]+");
1540 .
Case(
"fadd", Intrinsic::vector_reduce_fadd)
1541 .
Case(
"fmul", Intrinsic::vector_reduce_fmul)
1546 auto Args =
F->getFunctionType()->params();
1548 {Args[V2 ? 1 : 0]});
1554 if (Name.consume_front(
"splice"))
1558 if (Name.consume_front(
"experimental.stepvector.")) {
1562 F->getParent(),
ID,
F->getFunctionType()->getReturnType());
1567 if (Name.starts_with(
"flt.rounds")) {
1570 Intrinsic::get_rounding);
1575 if (Name.starts_with(
"invariant.group.barrier")) {
1577 auto Args =
F->getFunctionType()->params();
1578 Type* ObjectPtr[1] = {Args[0]};
1581 F->getParent(), Intrinsic::launder_invariant_group, ObjectPtr);
1586 bool IsLifetimeStart = Name.consume_front(
"lifetime.start");
1587 bool IsLifetimeEnd = !IsLifetimeStart && Name.consume_front(
"lifetime.end");
1588 if (IsLifetimeStart || IsLifetimeEnd) {
1589 if (
F->arg_size() == 2) {
1590 Intrinsic::ID IID = IsLifetimeStart ? Intrinsic::lifetime_start
1591 : Intrinsic::lifetime_end;
1596 F->getArg(1)->getType());
1598 }
else if (
F->arg_size() == 1 && Name ==
".i64") {
1618 .StartsWith(
"memcpy.", Intrinsic::memcpy)
1619 .StartsWith(
"memmove.", Intrinsic::memmove)
1621 if (
F->arg_size() == 5) {
1625 F->getFunctionType()->params().slice(0, 3);
1631 if (Name.starts_with(
"memset.") &&
F->arg_size() == 5) {
1634 const auto *FT =
F->getFunctionType();
1635 Type *ParamTypes[2] = {
1636 FT->getParamType(0),
1640 Intrinsic::memset, ParamTypes);
1646 .
StartsWith(
"masked.load", Intrinsic::masked_load)
1647 .
StartsWith(
"masked.gather", Intrinsic::masked_gather)
1648 .
StartsWith(
"masked.store", Intrinsic::masked_store)
1649 .
StartsWith(
"masked.scatter", Intrinsic::masked_scatter)
1651 if (MaskedID &&
F->arg_size() == 4) {
1653 if (MaskedID == Intrinsic::masked_load ||
1654 MaskedID == Intrinsic::masked_gather) {
1656 F->getParent(), MaskedID,
1657 {F->getReturnType(), F->getArg(0)->getType()});
1661 F->getParent(), MaskedID,
1662 {F->getArg(0)->getType(), F->getArg(1)->getType()});
1668 if (Name.consume_front(
"nvvm.")) {
1670 if (
F->arg_size() == 1) {
1673 .
Cases({
"brev32",
"brev64"}, Intrinsic::bitreverse)
1674 .Case(
"clz.i", Intrinsic::ctlz)
1675 .
Case(
"popc.i", Intrinsic::ctpop)
1679 {F->getReturnType()});
1682 }
else if (
F->arg_size() == 2) {
1685 .
Cases({
"max.s",
"max.i",
"max.ll"}, Intrinsic::smax)
1686 .Cases({
"min.s",
"min.i",
"min.ll"}, Intrinsic::smin)
1687 .Cases({
"max.us",
"max.ui",
"max.ull"}, Intrinsic::umax)
1688 .Cases({
"min.us",
"min.ui",
"min.ull"}, Intrinsic::umin)
1692 {F->getReturnType()});
1698 if (!
F->getReturnType()->getScalarType()->isBFloatTy()) {
1726 bool Expand =
false;
1727 if (Name.consume_front(
"abs."))
1730 Name ==
"i" || Name ==
"ll" || Name ==
"bf16" || Name ==
"bf16x2";
1731 else if (Name.consume_front(
"fabs."))
1733 Expand = Name ==
"f" || Name ==
"ftz.f" || Name ==
"d";
1734 else if (Name.consume_front(
"ex2.approx."))
1737 Name ==
"f" || Name ==
"ftz.f" || Name ==
"d" || Name ==
"f16x2";
1738 else if (Name.consume_front(
"atomic.load."))
1747 else if (Name.consume_front(
"atomic."))
1762 else if (Name.consume_front(
"bitcast."))
1765 Name ==
"f2i" || Name ==
"i2f" || Name ==
"ll2d" || Name ==
"d2ll";
1766 else if (Name.consume_front(
"rotate."))
1768 Expand = Name ==
"b32" || Name ==
"b64" || Name ==
"right.b64";
1769 else if (Name.consume_front(
"ptr.gen.to."))
1772 else if (Name.consume_front(
"ptr."))
1775 else if (Name.consume_front(
"ldg.global."))
1777 Expand = (Name.starts_with(
"i.") || Name.starts_with(
"f.") ||
1778 Name.starts_with(
"p."));
1781 .
Case(
"barrier0",
true)
1782 .
Case(
"barrier.n",
true)
1783 .
Case(
"barrier.sync.cnt",
true)
1784 .
Case(
"barrier.sync",
true)
1785 .
Case(
"barrier",
true)
1786 .
Case(
"bar.sync",
true)
1787 .
Case(
"barrier0.popc",
true)
1788 .
Case(
"barrier0.and",
true)
1789 .
Case(
"barrier0.or",
true)
1790 .
Case(
"clz.ll",
true)
1791 .
Case(
"popc.ll",
true)
1793 .
Case(
"swap.lo.hi.b64",
true)
1794 .
Case(
"tanh.approx.f32",
true)
1806 if (Name.starts_with(
"objectsize.")) {
1807 Type *Tys[2] = {
F->getReturnType(),
F->arg_begin()->getType() };
1808 if (
F->arg_size() == 2 ||
F->arg_size() == 3) {
1811 Intrinsic::objectsize, Tys);
1818 if (Name.starts_with(
"ptr.annotation.") &&
F->arg_size() == 4) {
1821 F->getParent(), Intrinsic::ptr_annotation,
1822 {F->arg_begin()->getType(), F->getArg(1)->getType()});
1828 if (Name.consume_front(
"riscv.")) {
1831 .
Case(
"aes32dsi", Intrinsic::riscv_aes32dsi)
1832 .
Case(
"aes32dsmi", Intrinsic::riscv_aes32dsmi)
1833 .
Case(
"aes32esi", Intrinsic::riscv_aes32esi)
1834 .
Case(
"aes32esmi", Intrinsic::riscv_aes32esmi)
1837 if (!
F->getFunctionType()->getParamType(2)->isIntegerTy(32)) {
1850 if (!
F->getFunctionType()->getParamType(2)->isIntegerTy(32) ||
1851 F->getFunctionType()->getReturnType()->isIntegerTy(64)) {
1860 .
StartsWith(
"sha256sig0", Intrinsic::riscv_sha256sig0)
1861 .
StartsWith(
"sha256sig1", Intrinsic::riscv_sha256sig1)
1862 .
StartsWith(
"sha256sum0", Intrinsic::riscv_sha256sum0)
1863 .
StartsWith(
"sha256sum1", Intrinsic::riscv_sha256sum1)
1868 if (
F->getFunctionType()->getReturnType()->isIntegerTy(64)) {
1877 if (Name ==
"clmul.i32" || Name ==
"clmul.i64") {
1879 F->getParent(), Intrinsic::clmul, {F->getReturnType()});
1888 if (Name ==
"stackprotectorcheck") {
1895 if (Name ==
"thread.pointer") {
1897 F->getParent(), Intrinsic::thread_pointer,
F->getReturnType());
1903 if (Name ==
"var.annotation" &&
F->arg_size() == 4) {
1906 F->getParent(), Intrinsic::var_annotation,
1907 {{F->arg_begin()->getType(), F->getArg(1)->getType()}});
1910 if (Name.consume_front(
"vector.splice")) {
1911 if (Name.starts_with(
".left") || Name.starts_with(
".right"))
1919 if (Name.consume_front(
"wasm.")) {
1922 .
StartsWith(
"fma.", Intrinsic::wasm_relaxed_madd)
1923 .
StartsWith(
"fms.", Intrinsic::wasm_relaxed_nmadd)
1924 .
StartsWith(
"laneselect.", Intrinsic::wasm_relaxed_laneselect)
1929 F->getReturnType());
1933 if (Name.consume_front(
"dot.i8x16.i7x16.")) {
1935 .
Case(
"signed", Intrinsic::wasm_relaxed_dot_i8x16_i7x16_signed)
1937 Intrinsic::wasm_relaxed_dot_i8x16_i7x16_add_signed)
1956 if (ST && (!
ST->isLiteral() ||
ST->isPacked()) &&
1966 std::string
Name =
F->getName().str();
1969 Name,
F->getParent());
1980 if (Result != std::nullopt) {
1996 bool CanUpgradeDebugIntrinsicsToRecords) {
2016 GV->
getName() ==
"llvm.global_dtors")) ||
2031 unsigned N =
Init->getNumOperands();
2032 std::vector<Constant *> NewCtors(
N);
2033 for (
unsigned i = 0; i !=
N; ++i) {
2036 Ctor->getAggregateElement(1),
2050 unsigned NumElts = ResultTy->getNumElements() * 8;
2054 Op = Builder.CreateBitCast(
Op, VecTy,
"cast");
2064 for (
unsigned l = 0; l != NumElts; l += 16)
2065 for (
unsigned i = 0; i != 16; ++i) {
2066 unsigned Idx = NumElts + i - Shift;
2068 Idx -= NumElts - 16;
2069 Idxs[l + i] = Idx + l;
2072 Res = Builder.CreateShuffleVector(Res,
Op,
ArrayRef(Idxs, NumElts));
2076 return Builder.CreateBitCast(Res, ResultTy,
"cast");
2084 unsigned NumElts = ResultTy->getNumElements() * 8;
2088 Op = Builder.CreateBitCast(
Op, VecTy,
"cast");
2098 for (
unsigned l = 0; l != NumElts; l += 16)
2099 for (
unsigned i = 0; i != 16; ++i) {
2100 unsigned Idx = i + Shift;
2102 Idx += NumElts - 16;
2103 Idxs[l + i] = Idx + l;
2106 Res = Builder.CreateShuffleVector(
Op, Res,
ArrayRef(Idxs, NumElts));
2110 return Builder.CreateBitCast(Res, ResultTy,
"cast");
2118 Mask = Builder.CreateBitCast(Mask, MaskTy);
2124 for (
unsigned i = 0; i != NumElts; ++i)
2126 Mask = Builder.CreateShuffleVector(Mask, Mask,
ArrayRef(Indices, NumElts),
2137 if (
C->isAllOnesValue())
2142 return Builder.CreateSelect(Mask, Op0, Op1);
2149 if (
C->isAllOnesValue())
2153 Mask->getType()->getIntegerBitWidth());
2154 Mask = Builder.CreateBitCast(Mask, MaskTy);
2155 Mask = Builder.CreateExtractElement(Mask, (
uint64_t)0);
2156 return Builder.CreateSelect(Mask, Op0, Op1);
2169 assert((IsVALIGN || NumElts % 16 == 0) &&
"Illegal NumElts for PALIGNR!");
2170 assert((!IsVALIGN || NumElts <= 16) &&
"NumElts too large for VALIGN!");
2175 ShiftVal &= (NumElts - 1);
2184 if (ShiftVal > 16) {
2192 for (
unsigned l = 0; l < NumElts; l += 16) {
2193 for (
unsigned i = 0; i != 16; ++i) {
2194 unsigned Idx = ShiftVal + i;
2195 if (!IsVALIGN && Idx >= 16)
2196 Idx += NumElts - 16;
2197 Indices[l + i] = Idx + l;
2202 Op1, Op0,
ArrayRef(Indices, NumElts),
"palignr");
2208 bool ZeroMask,
bool IndexForm) {
2211 unsigned EltWidth = Ty->getScalarSizeInBits();
2212 bool IsFloat = Ty->isFPOrFPVectorTy();
2214 if (VecWidth == 128 && EltWidth == 32 && IsFloat)
2215 IID = Intrinsic::x86_avx512_vpermi2var_ps_128;
2216 else if (VecWidth == 128 && EltWidth == 32 && !IsFloat)
2217 IID = Intrinsic::x86_avx512_vpermi2var_d_128;
2218 else if (VecWidth == 128 && EltWidth == 64 && IsFloat)
2219 IID = Intrinsic::x86_avx512_vpermi2var_pd_128;
2220 else if (VecWidth == 128 && EltWidth == 64 && !IsFloat)
2221 IID = Intrinsic::x86_avx512_vpermi2var_q_128;
2222 else if (VecWidth == 256 && EltWidth == 32 && IsFloat)
2223 IID = Intrinsic::x86_avx512_vpermi2var_ps_256;
2224 else if (VecWidth == 256 && EltWidth == 32 && !IsFloat)
2225 IID = Intrinsic::x86_avx512_vpermi2var_d_256;
2226 else if (VecWidth == 256 && EltWidth == 64 && IsFloat)
2227 IID = Intrinsic::x86_avx512_vpermi2var_pd_256;
2228 else if (VecWidth == 256 && EltWidth == 64 && !IsFloat)
2229 IID = Intrinsic::x86_avx512_vpermi2var_q_256;
2230 else if (VecWidth == 512 && EltWidth == 32 && IsFloat)
2231 IID = Intrinsic::x86_avx512_vpermi2var_ps_512;
2232 else if (VecWidth == 512 && EltWidth == 32 && !IsFloat)
2233 IID = Intrinsic::x86_avx512_vpermi2var_d_512;
2234 else if (VecWidth == 512 && EltWidth == 64 && IsFloat)
2235 IID = Intrinsic::x86_avx512_vpermi2var_pd_512;
2236 else if (VecWidth == 512 && EltWidth == 64 && !IsFloat)
2237 IID = Intrinsic::x86_avx512_vpermi2var_q_512;
2238 else if (VecWidth == 128 && EltWidth == 16)
2239 IID = Intrinsic::x86_avx512_vpermi2var_hi_128;
2240 else if (VecWidth == 256 && EltWidth == 16)
2241 IID = Intrinsic::x86_avx512_vpermi2var_hi_256;
2242 else if (VecWidth == 512 && EltWidth == 16)
2243 IID = Intrinsic::x86_avx512_vpermi2var_hi_512;
2244 else if (VecWidth == 128 && EltWidth == 8)
2245 IID = Intrinsic::x86_avx512_vpermi2var_qi_128;
2246 else if (VecWidth == 256 && EltWidth == 8)
2247 IID = Intrinsic::x86_avx512_vpermi2var_qi_256;
2248 else if (VecWidth == 512 && EltWidth == 8)
2249 IID = Intrinsic::x86_avx512_vpermi2var_qi_512;
2260 Value *V = Builder.CreateIntrinsic(IID, Args);
2272 Value *Res = Builder.CreateIntrinsic(IID, Ty, {Op0, Op1});
2283 bool IsRotateRight) {
2293 Amt = Builder.CreateIntCast(Amt, Ty->getScalarType(),
false);
2294 Amt = Builder.CreateVectorSplat(NumElts, Amt);
2297 Intrinsic::ID IID = IsRotateRight ? Intrinsic::fshr : Intrinsic::fshl;
2298 Value *Res = Builder.CreateIntrinsic(IID, Ty, {Src, Src, Amt});
2343 Value *Ext = Builder.CreateSExt(Cmp, Ty);
2348 bool IsShiftRight,
bool ZeroMask) {
2362 Amt = Builder.CreateIntCast(Amt, Ty->getScalarType(),
false);
2363 Amt = Builder.CreateVectorSplat(NumElts, Amt);
2366 Intrinsic::ID IID = IsShiftRight ? Intrinsic::fshr : Intrinsic::fshl;
2367 Value *Res = Builder.CreateIntrinsic(IID, Ty, {Op0, Op1, Amt});
2382 const Align Alignment =
2384 ?
Align(
Data->getType()->getPrimitiveSizeInBits().getFixedValue() / 8)
2389 if (
C->isAllOnesValue())
2390 return Builder.CreateAlignedStore(
Data, Ptr, Alignment);
2395 return Builder.CreateMaskedStore(
Data, Ptr, Alignment, Mask);
2401 const Align Alignment =
2410 if (
C->isAllOnesValue())
2411 return Builder.CreateAlignedLoad(ValTy, Ptr, Alignment);
2416 return Builder.CreateMaskedLoad(ValTy, Ptr, Alignment, Mask, Passthru);
2422 Value *Res = Builder.CreateIntrinsic(Intrinsic::abs, Ty,
2423 {Op0, Builder.getInt1(
false)});
2438 Constant *ShiftAmt = ConstantInt::get(Ty, 32);
2439 LHS = Builder.CreateShl(
LHS, ShiftAmt);
2440 LHS = Builder.CreateAShr(
LHS, ShiftAmt);
2441 RHS = Builder.CreateShl(
RHS, ShiftAmt);
2442 RHS = Builder.CreateAShr(
RHS, ShiftAmt);
2445 Constant *Mask = ConstantInt::get(Ty, 0xffffffff);
2446 LHS = Builder.CreateAnd(
LHS, Mask);
2447 RHS = Builder.CreateAnd(
RHS, Mask);
2464 if (!
C || !
C->isAllOnesValue())
2465 Vec = Builder.CreateAnd(Vec,
getX86MaskVec(Builder, Mask, NumElts));
2470 for (
unsigned i = 0; i != NumElts; ++i)
2472 for (
unsigned i = NumElts; i != 8; ++i)
2473 Indices[i] = NumElts + i % NumElts;
2474 Vec = Builder.CreateShuffleVector(Vec,
2478 return Builder.CreateBitCast(Vec, Builder.getIntNTy(std::max(NumElts, 8U)));
2482 unsigned CC,
bool Signed) {
2490 }
else if (CC == 7) {
2526 Value* AndNode = Builder.CreateAnd(Mask,
APInt(8, 1));
2527 Value* Cmp = Builder.CreateIsNotNull(AndNode);
2529 Value* Extract2 = Builder.CreateExtractElement(Src, (
uint64_t)0);
2530 Value*
Select = Builder.CreateSelect(Cmp, Extract1, Extract2);
2539 return Builder.CreateSExt(Mask, ReturnOp,
"vpmovm2");
2545 Name = Name.substr(12);
2550 if (Name.starts_with(
"max.p")) {
2551 if (VecWidth == 128 && EltWidth == 32)
2552 IID = Intrinsic::x86_sse_max_ps;
2553 else if (VecWidth == 128 && EltWidth == 64)
2554 IID = Intrinsic::x86_sse2_max_pd;
2555 else if (VecWidth == 256 && EltWidth == 32)
2556 IID = Intrinsic::x86_avx_max_ps_256;
2557 else if (VecWidth == 256 && EltWidth == 64)
2558 IID = Intrinsic::x86_avx_max_pd_256;
2561 }
else if (Name.starts_with(
"min.p")) {
2562 if (VecWidth == 128 && EltWidth == 32)
2563 IID = Intrinsic::x86_sse_min_ps;
2564 else if (VecWidth == 128 && EltWidth == 64)
2565 IID = Intrinsic::x86_sse2_min_pd;
2566 else if (VecWidth == 256 && EltWidth == 32)
2567 IID = Intrinsic::x86_avx_min_ps_256;
2568 else if (VecWidth == 256 && EltWidth == 64)
2569 IID = Intrinsic::x86_avx_min_pd_256;
2572 }
else if (Name.starts_with(
"pshuf.b.")) {
2573 if (VecWidth == 128)
2574 IID = Intrinsic::x86_ssse3_pshuf_b_128;
2575 else if (VecWidth == 256)
2576 IID = Intrinsic::x86_avx2_pshuf_b;
2577 else if (VecWidth == 512)
2578 IID = Intrinsic::x86_avx512_pshuf_b_512;
2581 }
else if (Name.starts_with(
"pmul.hr.sw.")) {
2582 if (VecWidth == 128)
2583 IID = Intrinsic::x86_ssse3_pmul_hr_sw_128;
2584 else if (VecWidth == 256)
2585 IID = Intrinsic::x86_avx2_pmul_hr_sw;
2586 else if (VecWidth == 512)
2587 IID = Intrinsic::x86_avx512_pmul_hr_sw_512;
2590 }
else if (Name.starts_with(
"pmulh.w.")) {
2591 if (VecWidth == 128)
2592 IID = Intrinsic::x86_sse2_pmulh_w;
2593 else if (VecWidth == 256)
2594 IID = Intrinsic::x86_avx2_pmulh_w;
2595 else if (VecWidth == 512)
2596 IID = Intrinsic::x86_avx512_pmulh_w_512;
2599 }
else if (Name.starts_with(
"pmulhu.w.")) {
2600 if (VecWidth == 128)
2601 IID = Intrinsic::x86_sse2_pmulhu_w;
2602 else if (VecWidth == 256)
2603 IID = Intrinsic::x86_avx2_pmulhu_w;
2604 else if (VecWidth == 512)
2605 IID = Intrinsic::x86_avx512_pmulhu_w_512;
2608 }
else if (Name.starts_with(
"pmaddw.d.")) {
2609 if (VecWidth == 128)
2610 IID = Intrinsic::x86_sse2_pmadd_wd;
2611 else if (VecWidth == 256)
2612 IID = Intrinsic::x86_avx2_pmadd_wd;
2613 else if (VecWidth == 512)
2614 IID = Intrinsic::x86_avx512_pmaddw_d_512;
2617 }
else if (Name.starts_with(
"pmaddubs.w.")) {
2618 if (VecWidth == 128)
2619 IID = Intrinsic::x86_ssse3_pmadd_ub_sw_128;
2620 else if (VecWidth == 256)
2621 IID = Intrinsic::x86_avx2_pmadd_ub_sw;
2622 else if (VecWidth == 512)
2623 IID = Intrinsic::x86_avx512_pmaddubs_w_512;
2626 }
else if (Name.starts_with(
"packsswb.")) {
2627 if (VecWidth == 128)
2628 IID = Intrinsic::x86_sse2_packsswb_128;
2629 else if (VecWidth == 256)
2630 IID = Intrinsic::x86_avx2_packsswb;
2631 else if (VecWidth == 512)
2632 IID = Intrinsic::x86_avx512_packsswb_512;
2635 }
else if (Name.starts_with(
"packssdw.")) {
2636 if (VecWidth == 128)
2637 IID = Intrinsic::x86_sse2_packssdw_128;
2638 else if (VecWidth == 256)
2639 IID = Intrinsic::x86_avx2_packssdw;
2640 else if (VecWidth == 512)
2641 IID = Intrinsic::x86_avx512_packssdw_512;
2644 }
else if (Name.starts_with(
"packuswb.")) {
2645 if (VecWidth == 128)
2646 IID = Intrinsic::x86_sse2_packuswb_128;
2647 else if (VecWidth == 256)
2648 IID = Intrinsic::x86_avx2_packuswb;
2649 else if (VecWidth == 512)
2650 IID = Intrinsic::x86_avx512_packuswb_512;
2653 }
else if (Name.starts_with(
"packusdw.")) {
2654 if (VecWidth == 128)
2655 IID = Intrinsic::x86_sse41_packusdw;
2656 else if (VecWidth == 256)
2657 IID = Intrinsic::x86_avx2_packusdw;
2658 else if (VecWidth == 512)
2659 IID = Intrinsic::x86_avx512_packusdw_512;
2662 }
else if (Name.starts_with(
"vpermilvar.")) {
2663 if (VecWidth == 128 && EltWidth == 32)
2664 IID = Intrinsic::x86_avx_vpermilvar_ps;
2665 else if (VecWidth == 128 && EltWidth == 64)
2666 IID = Intrinsic::x86_avx_vpermilvar_pd;
2667 else if (VecWidth == 256 && EltWidth == 32)
2668 IID = Intrinsic::x86_avx_vpermilvar_ps_256;
2669 else if (VecWidth == 256 && EltWidth == 64)
2670 IID = Intrinsic::x86_avx_vpermilvar_pd_256;
2671 else if (VecWidth == 512 && EltWidth == 32)
2672 IID = Intrinsic::x86_avx512_vpermilvar_ps_512;
2673 else if (VecWidth == 512 && EltWidth == 64)
2674 IID = Intrinsic::x86_avx512_vpermilvar_pd_512;
2677 }
else if (Name ==
"cvtpd2dq.256") {
2678 IID = Intrinsic::x86_avx_cvt_pd2dq_256;
2679 }
else if (Name ==
"cvtpd2ps.256") {
2680 IID = Intrinsic::x86_avx_cvt_pd2_ps_256;
2681 }
else if (Name ==
"cvttpd2dq.256") {
2682 IID = Intrinsic::x86_avx_cvtt_pd2dq_256;
2683 }
else if (Name ==
"cvttps2dq.128") {
2684 IID = Intrinsic::x86_sse2_cvttps2dq;
2685 }
else if (Name ==
"cvttps2dq.256") {
2686 IID = Intrinsic::x86_avx_cvtt_ps2dq_256;
2687 }
else if (Name.starts_with(
"permvar.")) {
2689 if (VecWidth == 256 && EltWidth == 32 && IsFloat)
2690 IID = Intrinsic::x86_avx2_permps;
2691 else if (VecWidth == 256 && EltWidth == 32 && !IsFloat)
2692 IID = Intrinsic::x86_avx2_permd;
2693 else if (VecWidth == 256 && EltWidth == 64 && IsFloat)
2694 IID = Intrinsic::x86_avx512_permvar_df_256;
2695 else if (VecWidth == 256 && EltWidth == 64 && !IsFloat)
2696 IID = Intrinsic::x86_avx512_permvar_di_256;
2697 else if (VecWidth == 512 && EltWidth == 32 && IsFloat)
2698 IID = Intrinsic::x86_avx512_permvar_sf_512;
2699 else if (VecWidth == 512 && EltWidth == 32 && !IsFloat)
2700 IID = Intrinsic::x86_avx512_permvar_si_512;
2701 else if (VecWidth == 512 && EltWidth == 64 && IsFloat)
2702 IID = Intrinsic::x86_avx512_permvar_df_512;
2703 else if (VecWidth == 512 && EltWidth == 64 && !IsFloat)
2704 IID = Intrinsic::x86_avx512_permvar_di_512;
2705 else if (VecWidth == 128 && EltWidth == 16)
2706 IID = Intrinsic::x86_avx512_permvar_hi_128;
2707 else if (VecWidth == 256 && EltWidth == 16)
2708 IID = Intrinsic::x86_avx512_permvar_hi_256;
2709 else if (VecWidth == 512 && EltWidth == 16)
2710 IID = Intrinsic::x86_avx512_permvar_hi_512;
2711 else if (VecWidth == 128 && EltWidth == 8)
2712 IID = Intrinsic::x86_avx512_permvar_qi_128;
2713 else if (VecWidth == 256 && EltWidth == 8)
2714 IID = Intrinsic::x86_avx512_permvar_qi_256;
2715 else if (VecWidth == 512 && EltWidth == 8)
2716 IID = Intrinsic::x86_avx512_permvar_qi_512;
2719 }
else if (Name.starts_with(
"dbpsadbw.")) {
2720 if (VecWidth == 128)
2721 IID = Intrinsic::x86_avx512_dbpsadbw_128;
2722 else if (VecWidth == 256)
2723 IID = Intrinsic::x86_avx512_dbpsadbw_256;
2724 else if (VecWidth == 512)
2725 IID = Intrinsic::x86_avx512_dbpsadbw_512;
2728 }
else if (Name.starts_with(
"pmultishift.qb.")) {
2729 if (VecWidth == 128)
2730 IID = Intrinsic::x86_avx512_pmultishift_qb_128;
2731 else if (VecWidth == 256)
2732 IID = Intrinsic::x86_avx512_pmultishift_qb_256;
2733 else if (VecWidth == 512)
2734 IID = Intrinsic::x86_avx512_pmultishift_qb_512;
2737 }
else if (Name.starts_with(
"conflict.")) {
2738 if (Name[9] ==
'd' && VecWidth == 128)
2739 IID = Intrinsic::x86_avx512_conflict_d_128;
2740 else if (Name[9] ==
'd' && VecWidth == 256)
2741 IID = Intrinsic::x86_avx512_conflict_d_256;
2742 else if (Name[9] ==
'd' && VecWidth == 512)
2743 IID = Intrinsic::x86_avx512_conflict_d_512;
2744 else if (Name[9] ==
'q' && VecWidth == 128)
2745 IID = Intrinsic::x86_avx512_conflict_q_128;
2746 else if (Name[9] ==
'q' && VecWidth == 256)
2747 IID = Intrinsic::x86_avx512_conflict_q_256;
2748 else if (Name[9] ==
'q' && VecWidth == 512)
2749 IID = Intrinsic::x86_avx512_conflict_q_512;
2752 }
else if (Name.starts_with(
"pavg.")) {
2753 if (Name[5] ==
'b' && VecWidth == 128)
2754 IID = Intrinsic::x86_sse2_pavg_b;
2755 else if (Name[5] ==
'b' && VecWidth == 256)
2756 IID = Intrinsic::x86_avx2_pavg_b;
2757 else if (Name[5] ==
'b' && VecWidth == 512)
2758 IID = Intrinsic::x86_avx512_pavg_b_512;
2759 else if (Name[5] ==
'w' && VecWidth == 128)
2760 IID = Intrinsic::x86_sse2_pavg_w;
2761 else if (Name[5] ==
'w' && VecWidth == 256)
2762 IID = Intrinsic::x86_avx2_pavg_w;
2763 else if (Name[5] ==
'w' && VecWidth == 512)
2764 IID = Intrinsic::x86_avx512_pavg_w_512;
2773 Rep = Builder.CreateIntrinsic(IID, Args);
2784 if (AsmStr->find(
"mov\tfp") == 0 &&
2785 AsmStr->find(
"objc_retainAutoreleaseReturnValue") != std::string::npos &&
2786 (Pos = AsmStr->find(
"# marker")) != std::string::npos) {
2787 AsmStr->replace(Pos, 1,
";");
2793 Value *Rep =
nullptr;
2795 if (Name ==
"abs.i" || Name ==
"abs.ll") {
2797 Rep = Builder.CreateIntrinsic(Intrinsic::abs, {Arg->
getType()},
2798 {Arg, Builder.getTrue()},
2800 }
else if (Name ==
"abs.bf16" || Name ==
"abs.bf16x2") {
2801 Type *Ty = (Name ==
"abs.bf16")
2805 Value *Abs = Builder.CreateUnaryIntrinsic(Intrinsic::nvvm_fabs, Arg);
2806 Rep = Builder.CreateBitCast(Abs, CI->
getType());
2807 }
else if (Name ==
"fabs.f" || Name ==
"fabs.ftz.f" || Name ==
"fabs.d") {
2808 Intrinsic::ID IID = (Name ==
"fabs.ftz.f") ? Intrinsic::nvvm_fabs_ftz
2809 : Intrinsic::nvvm_fabs;
2810 Rep = Builder.CreateUnaryIntrinsic(IID, CI->
getArgOperand(0));
2811 }
else if (Name.consume_front(
"ex2.approx.")) {
2813 Intrinsic::ID IID = Name.starts_with(
"ftz") ? Intrinsic::nvvm_ex2_approx_ftz
2814 : Intrinsic::nvvm_ex2_approx;
2815 Rep = Builder.CreateUnaryIntrinsic(IID, CI->
getArgOperand(0));
2816 }
else if (Name.starts_with(
"atomic.load.add.f32.p") ||
2817 Name.starts_with(
"atomic.load.add.f64.p")) {
2820 Rep = Builder.CreateAtomicRMW(
2826 }
else if (Name.starts_with(
"atomic.load.inc.32.p") ||
2827 Name.starts_with(
"atomic.load.dec.32.p")) {
2832 Rep = Builder.CreateAtomicRMW(
2836 }
else if (Name.starts_with(
"atomic.") && Name.contains(
".gen.")) {
2842 Op.contains(
".cta.") ?
"block" :
"");
2843 if (
Op.starts_with(
"cas.")) {
2845 Value *Pair = Builder.CreateAtomicCmpXchg(
2848 Rep = Builder.CreateExtractValue(Pair, 0);
2866 "unexpected nvvm scoped atomic intrinsic");
2867 Rep = Builder.CreateAtomicRMW(BinOp, Ptr, Val,
MaybeAlign(),
2870 }
else if (Name ==
"clz.ll") {
2873 Value *Ctlz = Builder.CreateIntrinsic(Intrinsic::ctlz, {Arg->
getType()},
2874 {Arg, Builder.getFalse()},
2876 Rep = Builder.CreateTrunc(Ctlz, Builder.getInt32Ty(),
"ctlz.trunc");
2877 }
else if (Name ==
"popc.ll") {
2881 Value *Popc = Builder.CreateIntrinsic(Intrinsic::ctpop, {Arg->
getType()},
2882 Arg,
nullptr,
"ctpop");
2883 Rep = Builder.CreateTrunc(Popc, Builder.getInt32Ty(),
"ctpop.trunc");
2884 }
else if (Name ==
"h2f") {
2886 Builder.CreateBitCast(CI->
getArgOperand(0), Builder.getHalfTy());
2887 Rep = Builder.CreateFPExt(Cast, Builder.getFloatTy());
2888 }
else if (Name.consume_front(
"bitcast.") &&
2889 (Name ==
"f2i" || Name ==
"i2f" || Name ==
"ll2d" ||
2892 }
else if (Name ==
"rotate.b32") {
2895 Rep = Builder.CreateIntrinsic(Builder.getInt32Ty(), Intrinsic::fshl,
2896 {Arg, Arg, ShiftAmt});
2897 }
else if (Name ==
"rotate.b64") {
2901 Rep = Builder.CreateIntrinsic(Int64Ty, Intrinsic::fshl,
2902 {Arg, Arg, ZExtShiftAmt});
2903 }
else if (Name ==
"rotate.right.b64") {
2907 Rep = Builder.CreateIntrinsic(Int64Ty, Intrinsic::fshr,
2908 {Arg, Arg, ZExtShiftAmt});
2909 }
else if (Name ==
"swap.lo.hi.b64") {
2912 Rep = Builder.CreateIntrinsic(Int64Ty, Intrinsic::fshl,
2913 {Arg, Arg, Builder.getInt64(32)});
2914 }
else if ((Name.consume_front(
"ptr.gen.to.") &&
2917 Name.starts_with(
".to.gen"))) {
2919 }
else if (Name.consume_front(
"ldg.global")) {
2923 Value *ASC = Builder.CreateAddrSpaceCast(Ptr, Builder.getPtrTy(1));
2926 LD->setMetadata(LLVMContext::MD_invariant_load, MD);
2928 }
else if (Name ==
"tanh.approx.f32") {
2932 Rep = Builder.CreateUnaryIntrinsic(Intrinsic::tanh, CI->
getArgOperand(0),
2934 }
else if (Name ==
"barrier0" || Name ==
"barrier.n" || Name ==
"bar.sync") {
2936 Name.ends_with(
'0') ? Builder.getInt32(0) : CI->
getArgOperand(0);
2937 Rep = Builder.CreateIntrinsic(Intrinsic::nvvm_barrier_cta_sync_aligned_all,
2939 }
else if (Name ==
"barrier") {
2940 Rep = Builder.CreateIntrinsic(
2941 Intrinsic::nvvm_barrier_cta_sync_aligned_count, {},
2943 }
else if (Name ==
"barrier.sync") {
2944 Rep = Builder.CreateIntrinsic(Intrinsic::nvvm_barrier_cta_sync_all, {},
2946 }
else if (Name ==
"barrier.sync.cnt") {
2947 Rep = Builder.CreateIntrinsic(Intrinsic::nvvm_barrier_cta_sync_count, {},
2949 }
else if (Name ==
"barrier0.popc" || Name ==
"barrier0.and" ||
2950 Name ==
"barrier0.or") {
2952 C = Builder.CreateICmpNE(
C, Builder.getInt32(0));
2956 .
Case(
"barrier0.popc",
2957 Intrinsic::nvvm_barrier_cta_red_popc_aligned_all)
2958 .
Case(
"barrier0.and",
2959 Intrinsic::nvvm_barrier_cta_red_and_aligned_all)
2960 .
Case(
"barrier0.or",
2961 Intrinsic::nvvm_barrier_cta_red_or_aligned_all);
2962 Value *Bar = Builder.CreateIntrinsic(IID, {}, {Builder.getInt32(0),
C});
2963 Rep = Builder.CreateZExt(Bar, CI->
getType());
2967 !
F->getReturnType()->getScalarType()->isBFloatTy()) {
2977 ? Builder.CreateBitCast(Arg, NewType)
2980 Rep = Builder.CreateCall(NewFn, Args);
2981 if (
F->getReturnType()->isIntegerTy())
2982 Rep = Builder.CreateBitCast(Rep,
F->getReturnType());
2992 Value *Rep =
nullptr;
2994 if (Name.starts_with(
"sse4a.movnt.")) {
3006 Builder.CreateExtractElement(Arg1, (
uint64_t)0,
"extractelement");
3009 SI->setMetadata(LLVMContext::MD_nontemporal,
Node);
3010 }
else if (Name.starts_with(
"avx.movnt.") ||
3011 Name.starts_with(
"avx512.storent.")) {
3023 SI->setMetadata(LLVMContext::MD_nontemporal,
Node);
3024 }
else if (Name ==
"sse2.storel.dq") {
3029 Value *BC0 = Builder.CreateBitCast(Arg1, NewVecTy,
"cast");
3030 Value *Elt = Builder.CreateExtractElement(BC0, (
uint64_t)0);
3031 Builder.CreateAlignedStore(Elt, Arg0,
Align(1));
3032 }
else if (Name.starts_with(
"sse.storeu.") ||
3033 Name.starts_with(
"sse2.storeu.") ||
3034 Name.starts_with(
"avx.storeu.")) {
3037 Builder.CreateAlignedStore(Arg1, Arg0,
Align(1));
3038 }
else if (Name ==
"avx512.mask.store.ss") {
3042 }
else if (Name.starts_with(
"avx512.mask.store")) {
3044 bool Aligned = Name[17] !=
'u';
3047 }
else if (Name.starts_with(
"sse2.pcmp") || Name.starts_with(
"avx2.pcmp")) {
3050 bool CmpEq = Name[9] ==
'e';
3053 Rep = Builder.CreateSExt(Rep, CI->
getType(),
"");
3054 }
else if (Name.starts_with(
"avx512.broadcastm")) {
3061 Rep = Builder.CreateVectorSplat(NumElts, Rep);
3062 }
else if (Name ==
"sse.sqrt.ss" || Name ==
"sse2.sqrt.sd") {
3064 Value *Elt0 = Builder.CreateExtractElement(Vec, (
uint64_t)0);
3065 Elt0 = Builder.CreateIntrinsic(Intrinsic::sqrt, Elt0->
getType(), Elt0);
3066 Rep = Builder.CreateInsertElement(Vec, Elt0, (
uint64_t)0);
3067 }
else if (Name.starts_with(
"avx.sqrt.p") ||
3068 Name.starts_with(
"sse2.sqrt.p") ||
3069 Name.starts_with(
"sse.sqrt.p")) {
3070 Rep = Builder.CreateIntrinsic(Intrinsic::sqrt, CI->
getType(),
3071 {CI->getArgOperand(0)});
3072 }
else if (Name.starts_with(
"avx512.mask.sqrt.p")) {
3076 Intrinsic::ID IID = Name[18] ==
's' ? Intrinsic::x86_avx512_sqrt_ps_512
3077 : Intrinsic::x86_avx512_sqrt_pd_512;
3080 Rep = Builder.CreateIntrinsic(IID, Args);
3082 Rep = Builder.CreateIntrinsic(Intrinsic::sqrt, CI->
getType(),
3083 {CI->getArgOperand(0)});
3087 }
else if (Name.starts_with(
"avx512.ptestm") ||
3088 Name.starts_with(
"avx512.ptestnm")) {
3092 Rep = Builder.CreateAnd(Op0, Op1);
3098 Rep = Builder.CreateICmp(Pred, Rep, Zero);
3100 }
else if (Name.starts_with(
"avx512.mask.pbroadcast")) {
3103 Rep = Builder.CreateVectorSplat(NumElts, CI->
getArgOperand(0));
3106 }
else if (Name.starts_with(
"avx512.kunpck")) {
3111 for (
unsigned i = 0; i != NumElts; ++i)
3120 Rep = Builder.CreateShuffleVector(
RHS,
LHS,
ArrayRef(Indices, NumElts));
3121 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3122 }
else if (Name ==
"avx512.kand.w") {
3125 Rep = Builder.CreateAnd(
LHS,
RHS);
3126 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3127 }
else if (Name ==
"avx512.kandn.w") {
3130 LHS = Builder.CreateNot(
LHS);
3131 Rep = Builder.CreateAnd(
LHS,
RHS);
3132 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3133 }
else if (Name ==
"avx512.kor.w") {
3136 Rep = Builder.CreateOr(
LHS,
RHS);
3137 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3138 }
else if (Name ==
"avx512.kxor.w") {
3141 Rep = Builder.CreateXor(
LHS,
RHS);
3142 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3143 }
else if (Name ==
"avx512.kxnor.w") {
3146 LHS = Builder.CreateNot(
LHS);
3147 Rep = Builder.CreateXor(
LHS,
RHS);
3148 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3149 }
else if (Name ==
"avx512.knot.w") {
3151 Rep = Builder.CreateNot(Rep);
3152 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3153 }
else if (Name ==
"avx512.kortestz.w" || Name ==
"avx512.kortestc.w") {
3156 Rep = Builder.CreateOr(
LHS,
RHS);
3157 Rep = Builder.CreateBitCast(Rep, Builder.getInt16Ty());
3159 if (Name[14] ==
'c')
3163 Rep = Builder.CreateICmpEQ(Rep,
C);
3164 Rep = Builder.CreateZExt(Rep, Builder.getInt32Ty());
3165 }
else if (Name ==
"sse.add.ss" || Name ==
"sse2.add.sd" ||
3166 Name ==
"sse.sub.ss" || Name ==
"sse2.sub.sd" ||
3167 Name ==
"sse.mul.ss" || Name ==
"sse2.mul.sd" ||
3168 Name ==
"sse.div.ss" || Name ==
"sse2.div.sd") {
3171 ConstantInt::get(I32Ty, 0));
3173 ConstantInt::get(I32Ty, 0));
3175 if (Name.contains(
".add."))
3176 EltOp = Builder.CreateFAdd(Elt0, Elt1);
3177 else if (Name.contains(
".sub."))
3178 EltOp = Builder.CreateFSub(Elt0, Elt1);
3179 else if (Name.contains(
".mul."))
3180 EltOp = Builder.CreateFMul(Elt0, Elt1);
3182 EltOp = Builder.CreateFDiv(Elt0, Elt1);
3183 Rep = Builder.CreateInsertElement(CI->
getArgOperand(0), EltOp,
3184 ConstantInt::get(I32Ty, 0));
3185 }
else if (Name.starts_with(
"avx512.mask.pcmp")) {
3187 bool CmpEq = Name[16] ==
'e';
3189 }
else if (Name.starts_with(
"avx512.mask.vpshufbitqmb.")) {
3198 IID = Intrinsic::x86_avx512_vpshufbitqmb_128;
3201 IID = Intrinsic::x86_avx512_vpshufbitqmb_256;
3204 IID = Intrinsic::x86_avx512_vpshufbitqmb_512;
3211 }
else if (Name.starts_with(
"avx512.mask.fpclass.p")) {
3216 if (VecWidth == 128 && EltWidth == 32)
3217 IID = Intrinsic::x86_avx512_fpclass_ps_128;
3218 else if (VecWidth == 256 && EltWidth == 32)
3219 IID = Intrinsic::x86_avx512_fpclass_ps_256;
3220 else if (VecWidth == 512 && EltWidth == 32)
3221 IID = Intrinsic::x86_avx512_fpclass_ps_512;
3222 else if (VecWidth == 128 && EltWidth == 64)
3223 IID = Intrinsic::x86_avx512_fpclass_pd_128;
3224 else if (VecWidth == 256 && EltWidth == 64)
3225 IID = Intrinsic::x86_avx512_fpclass_pd_256;
3226 else if (VecWidth == 512 && EltWidth == 64)
3227 IID = Intrinsic::x86_avx512_fpclass_pd_512;
3234 }
else if (Name.starts_with(
"avx512.cmp.p")) {
3236 Type *OpTy = Args[0]->getType();
3240 if (VecWidth == 128 && EltWidth == 32)
3241 IID = Intrinsic::x86_avx512_mask_cmp_ps_128;
3242 else if (VecWidth == 256 && EltWidth == 32)
3243 IID = Intrinsic::x86_avx512_mask_cmp_ps_256;
3244 else if (VecWidth == 512 && EltWidth == 32)
3245 IID = Intrinsic::x86_avx512_mask_cmp_ps_512;
3246 else if (VecWidth == 128 && EltWidth == 64)
3247 IID = Intrinsic::x86_avx512_mask_cmp_pd_128;
3248 else if (VecWidth == 256 && EltWidth == 64)
3249 IID = Intrinsic::x86_avx512_mask_cmp_pd_256;
3250 else if (VecWidth == 512 && EltWidth == 64)
3251 IID = Intrinsic::x86_avx512_mask_cmp_pd_512;
3256 if (VecWidth == 512)
3258 Args.push_back(Mask);
3260 Rep = Builder.CreateIntrinsic(IID, Args);
3261 }
else if (Name.starts_with(
"avx512.mask.cmp.")) {
3265 }
else if (Name.starts_with(
"avx512.mask.ucmp.")) {
3268 }
else if (Name.starts_with(
"avx512.cvtb2mask.") ||
3269 Name.starts_with(
"avx512.cvtw2mask.") ||
3270 Name.starts_with(
"avx512.cvtd2mask.") ||
3271 Name.starts_with(
"avx512.cvtq2mask.")) {
3276 }
else if (Name ==
"ssse3.pabs.b.128" || Name ==
"ssse3.pabs.w.128" ||
3277 Name ==
"ssse3.pabs.d.128" || Name.starts_with(
"avx2.pabs") ||
3278 Name.starts_with(
"avx512.mask.pabs")) {
3280 }
else if (Name ==
"sse41.pmaxsb" || Name ==
"sse2.pmaxs.w" ||
3281 Name ==
"sse41.pmaxsd" || Name.starts_with(
"avx2.pmaxs") ||
3282 Name.starts_with(
"avx512.mask.pmaxs")) {
3284 }
else if (Name ==
"sse2.pmaxu.b" || Name ==
"sse41.pmaxuw" ||
3285 Name ==
"sse41.pmaxud" || Name.starts_with(
"avx2.pmaxu") ||
3286 Name.starts_with(
"avx512.mask.pmaxu")) {
3288 }
else if (Name ==
"sse41.pminsb" || Name ==
"sse2.pmins.w" ||
3289 Name ==
"sse41.pminsd" || Name.starts_with(
"avx2.pmins") ||
3290 Name.starts_with(
"avx512.mask.pmins")) {
3292 }
else if (Name ==
"sse2.pminu.b" || Name ==
"sse41.pminuw" ||
3293 Name ==
"sse41.pminud" || Name.starts_with(
"avx2.pminu") ||
3294 Name.starts_with(
"avx512.mask.pminu")) {
3296 }
else if (Name ==
"sse2.pmulu.dq" || Name ==
"avx2.pmulu.dq" ||
3297 Name ==
"avx512.pmulu.dq.512" ||
3298 Name.starts_with(
"avx512.mask.pmulu.dq.")) {
3300 }
else if (Name ==
"sse41.pmuldq" || Name ==
"avx2.pmul.dq" ||
3301 Name ==
"avx512.pmul.dq.512" ||
3302 Name.starts_with(
"avx512.mask.pmul.dq.")) {
3304 }
else if (Name ==
"sse.cvtsi2ss" || Name ==
"sse2.cvtsi2sd" ||
3305 Name ==
"sse.cvtsi642ss" || Name ==
"sse2.cvtsi642sd") {
3310 }
else if (Name ==
"avx512.cvtusi2sd") {
3315 }
else if (Name ==
"sse2.cvtss2sd") {
3317 Rep = Builder.CreateFPExt(
3320 }
else if (Name ==
"sse2.cvtdq2pd" || Name ==
"sse2.cvtdq2ps" ||
3321 Name ==
"avx.cvtdq2.pd.256" || Name ==
"avx.cvtdq2.ps.256" ||
3322 Name.starts_with(
"avx512.mask.cvtdq2pd.") ||
3323 Name.starts_with(
"avx512.mask.cvtudq2pd.") ||
3324 Name.starts_with(
"avx512.mask.cvtdq2ps.") ||
3325 Name.starts_with(
"avx512.mask.cvtudq2ps.") ||
3326 Name.starts_with(
"avx512.mask.cvtqq2pd.") ||
3327 Name.starts_with(
"avx512.mask.cvtuqq2pd.") ||
3328 Name ==
"avx512.mask.cvtqq2ps.256" ||
3329 Name ==
"avx512.mask.cvtqq2ps.512" ||
3330 Name ==
"avx512.mask.cvtuqq2ps.256" ||
3331 Name ==
"avx512.mask.cvtuqq2ps.512" || Name ==
"sse2.cvtps2pd" ||
3332 Name ==
"avx.cvt.ps2.pd.256" ||
3333 Name ==
"avx512.mask.cvtps2pd.128" ||
3334 Name ==
"avx512.mask.cvtps2pd.256") {
3339 unsigned NumDstElts = DstTy->getNumElements();
3341 assert(NumDstElts == 2 &&
"Unexpected vector size");
3342 Rep = Builder.CreateShuffleVector(Rep, Rep,
ArrayRef<int>{0, 1});
3345 bool IsPS2PD = SrcTy->getElementType()->isFloatTy();
3346 bool IsUnsigned = Name.contains(
"cvtu");
3348 Rep = Builder.CreateFPExt(Rep, DstTy,
"cvtps2pd");
3352 Intrinsic::ID IID = IsUnsigned ? Intrinsic::x86_avx512_uitofp_round
3353 : Intrinsic::x86_avx512_sitofp_round;
3354 Rep = Builder.CreateIntrinsic(IID, {DstTy, SrcTy},
3357 Rep = IsUnsigned ? Builder.CreateUIToFP(Rep, DstTy,
"cvt")
3358 : Builder.CreateSIToFP(Rep, DstTy,
"cvt");
3364 }
else if (Name.starts_with(
"avx512.mask.vcvtph2ps.") ||
3365 Name.starts_with(
"vcvtph2ps.")) {
3369 unsigned NumDstElts = DstTy->getNumElements();
3370 if (NumDstElts != SrcTy->getNumElements()) {
3371 assert(NumDstElts == 4 &&
"Unexpected vector size");
3372 Rep = Builder.CreateShuffleVector(Rep, Rep,
ArrayRef<int>{0, 1, 2, 3});
3374 Rep = Builder.CreateBitCast(
3376 Rep = Builder.CreateFPExt(Rep, DstTy,
"cvtph2ps");
3380 }
else if (Name.starts_with(
"avx512.mask.load")) {
3382 bool Aligned = Name[16] !=
'u';
3385 }
else if (Name.starts_with(
"avx512.mask.expand.load.")) {
3389 ResultTy->getNumElements());
3390 Rep = Builder.CreateIntrinsic(
3391 Intrinsic::masked_expandload, {ResultTy, PtrTy},
3393 }
else if (Name.starts_with(
"avx512.mask.compress.store.")) {
3399 Rep = Builder.CreateIntrinsic(
3400 Intrinsic::masked_compressstore, {ResultTy, PtrTy},
3402 }
else if (Name.starts_with(
"avx512.mask.compress.") ||
3403 Name.starts_with(
"avx512.mask.expand.")) {
3407 ResultTy->getNumElements());
3409 bool IsCompress = Name[12] ==
'c';
3410 Intrinsic::ID IID = IsCompress ? Intrinsic::x86_avx512_mask_compress
3411 : Intrinsic::x86_avx512_mask_expand;
3412 Rep = Builder.CreateIntrinsic(
3414 }
else if (Name.starts_with(
"xop.vpcom")) {
3416 if (Name.ends_with(
"ub") || Name.ends_with(
"uw") || Name.ends_with(
"ud") ||
3417 Name.ends_with(
"uq"))
3419 else if (Name.ends_with(
"b") || Name.ends_with(
"w") ||
3420 Name.ends_with(
"d") || Name.ends_with(
"q"))
3429 Name = Name.substr(9);
3430 if (Name.starts_with(
"lt"))
3432 else if (Name.starts_with(
"le"))
3434 else if (Name.starts_with(
"gt"))
3436 else if (Name.starts_with(
"ge"))
3438 else if (Name.starts_with(
"eq"))
3440 else if (Name.starts_with(
"ne"))
3442 else if (Name.starts_with(
"false"))
3444 else if (Name.starts_with(
"true"))
3451 }
else if (Name.starts_with(
"xop.vpcmov")) {
3453 Value *NotSel = Builder.CreateNot(Sel);
3456 Rep = Builder.CreateOr(Sel0, Sel1);
3457 }
else if (Name.starts_with(
"xop.vprot") || Name.starts_with(
"avx512.prol") ||
3458 Name.starts_with(
"avx512.mask.prol")) {
3460 }
else if (Name.starts_with(
"avx512.pror") ||
3461 Name.starts_with(
"avx512.mask.pror")) {
3463 }
else if (Name.starts_with(
"avx512.vpshld.") ||
3464 Name.starts_with(
"avx512.mask.vpshld") ||
3465 Name.starts_with(
"avx512.maskz.vpshld")) {
3466 bool ZeroMask = Name[11] ==
'z';
3468 }
else if (Name.starts_with(
"avx512.vpshrd.") ||
3469 Name.starts_with(
"avx512.mask.vpshrd") ||
3470 Name.starts_with(
"avx512.maskz.vpshrd")) {
3471 bool ZeroMask = Name[11] ==
'z';
3473 }
else if (Name ==
"sse42.crc32.64.8") {
3476 Rep = Builder.CreateIntrinsic(Intrinsic::x86_sse42_crc32_32_8,
3478 Rep = Builder.CreateZExt(Rep, CI->
getType(),
"");
3479 }
else if (Name.starts_with(
"avx.vbroadcast.s") ||
3480 Name.starts_with(
"avx512.vbroadcast.s")) {
3483 Type *EltTy = VecTy->getElementType();
3484 unsigned EltNum = VecTy->getNumElements();
3488 for (
unsigned I = 0;
I < EltNum; ++
I)
3489 Rep = Builder.CreateInsertElement(Rep, Load, ConstantInt::get(I32Ty,
I));
3490 }
else if (Name.starts_with(
"sse41.pmovsx") ||
3491 Name.starts_with(
"sse41.pmovzx") ||
3492 Name.starts_with(
"avx2.pmovsx") ||
3493 Name.starts_with(
"avx2.pmovzx") ||
3494 Name.starts_with(
"avx512.mask.pmovsx") ||
3495 Name.starts_with(
"avx512.mask.pmovzx")) {
3497 unsigned NumDstElts = DstTy->getNumElements();
3501 for (
unsigned i = 0; i != NumDstElts; ++i)
3506 bool DoSext = Name.contains(
"pmovsx");
3508 DoSext ? Builder.CreateSExt(SV, DstTy) : Builder.CreateZExt(SV, DstTy);
3513 }
else if (Name ==
"avx512.mask.pmov.qd.256" ||
3514 Name ==
"avx512.mask.pmov.qd.512" ||
3515 Name ==
"avx512.mask.pmov.wb.256" ||
3516 Name ==
"avx512.mask.pmov.wb.512") {
3521 }
else if (Name.starts_with(
"avx.vbroadcastf128") ||
3522 Name ==
"avx2.vbroadcasti128") {
3528 if (NumSrcElts == 2)
3529 Rep = Builder.CreateShuffleVector(Load,
ArrayRef<int>{0, 1, 0, 1});
3531 Rep = Builder.CreateShuffleVector(Load,
3533 }
else if (Name.starts_with(
"avx512.mask.shuf.i") ||
3534 Name.starts_with(
"avx512.mask.shuf.f")) {
3539 unsigned ControlBitsMask = NumLanes - 1;
3540 unsigned NumControlBits = NumLanes / 2;
3543 for (
unsigned l = 0; l != NumLanes; ++l) {
3544 unsigned LaneMask = (Imm >> (l * NumControlBits)) & ControlBitsMask;
3546 if (l >= NumLanes / 2)
3547 LaneMask += NumLanes;
3548 for (
unsigned i = 0; i != NumElementsInLane; ++i)
3549 ShuffleMask.push_back(LaneMask * NumElementsInLane + i);
3555 }
else if (Name.starts_with(
"avx512.mask.broadcastf") ||
3556 Name.starts_with(
"avx512.mask.broadcasti")) {
3559 unsigned NumDstElts =
3563 for (
unsigned i = 0; i != NumDstElts; ++i)
3564 ShuffleMask[i] = i % NumSrcElts;
3570 }
else if (Name.starts_with(
"avx2.pbroadcast") ||
3571 Name.starts_with(
"avx2.vbroadcast") ||
3572 Name.starts_with(
"avx512.pbroadcast") ||
3573 Name.starts_with(
"avx512.mask.broadcast.s")) {
3580 Rep = Builder.CreateShuffleVector(
Op, M);
3585 }
else if (Name.starts_with(
"sse2.padds.") ||
3586 Name.starts_with(
"avx2.padds.") ||
3587 Name.starts_with(
"avx512.padds.") ||
3588 Name.starts_with(
"avx512.mask.padds.")) {
3590 }
else if (Name.starts_with(
"sse2.psubs.") ||
3591 Name.starts_with(
"avx2.psubs.") ||
3592 Name.starts_with(
"avx512.psubs.") ||
3593 Name.starts_with(
"avx512.mask.psubs.")) {
3595 }
else if (Name.starts_with(
"sse2.paddus.") ||
3596 Name.starts_with(
"avx2.paddus.") ||
3597 Name.starts_with(
"avx512.mask.paddus.")) {
3599 }
else if (Name.starts_with(
"sse2.psubus.") ||
3600 Name.starts_with(
"avx2.psubus.") ||
3601 Name.starts_with(
"avx512.mask.psubus.")) {
3603 }
else if (Name.starts_with(
"avx512.mask.palignr.")) {
3608 }
else if (Name.starts_with(
"avx512.mask.valign.")) {
3612 }
else if (Name ==
"sse2.psll.dq" || Name ==
"avx2.psll.dq") {
3617 }
else if (Name ==
"sse2.psrl.dq" || Name ==
"avx2.psrl.dq") {
3622 }
else if (Name ==
"sse2.psll.dq.bs" || Name ==
"avx2.psll.dq.bs" ||
3623 Name ==
"avx512.psll.dq.512") {
3627 }
else if (Name ==
"sse2.psrl.dq.bs" || Name ==
"avx2.psrl.dq.bs" ||
3628 Name ==
"avx512.psrl.dq.512") {
3632 }
else if (Name ==
"sse41.pblendw" || Name.starts_with(
"sse41.blendp") ||
3633 Name.starts_with(
"avx.blend.p") || Name ==
"avx2.pblendw" ||
3634 Name.starts_with(
"avx2.pblendd.")) {
3639 unsigned NumElts = VecTy->getNumElements();
3642 for (
unsigned i = 0; i != NumElts; ++i)
3643 Idxs[i] = ((Imm >> (i % 8)) & 1) ? i + NumElts : i;
3645 Rep = Builder.CreateShuffleVector(Op0, Op1, Idxs);
3646 }
else if (Name.starts_with(
"avx.vinsertf128.") ||
3647 Name ==
"avx2.vinserti128" ||
3648 Name.starts_with(
"avx512.mask.insert")) {
3652 unsigned DstNumElts =
3654 unsigned SrcNumElts =
3656 unsigned Scale = DstNumElts / SrcNumElts;
3663 for (
unsigned i = 0; i != SrcNumElts; ++i)
3665 for (
unsigned i = SrcNumElts; i != DstNumElts; ++i)
3666 Idxs[i] = SrcNumElts;
3667 Rep = Builder.CreateShuffleVector(Op1, Idxs);
3681 for (
unsigned i = 0; i != DstNumElts; ++i)
3684 for (
unsigned i = 0; i != SrcNumElts; ++i)
3685 Idxs[i + Imm * SrcNumElts] = i + DstNumElts;
3686 Rep = Builder.CreateShuffleVector(Op0, Rep, Idxs);
3692 }
else if (Name.starts_with(
"avx.vextractf128.") ||
3693 Name ==
"avx2.vextracti128" ||
3694 Name.starts_with(
"avx512.mask.vextract")) {
3697 unsigned DstNumElts =
3699 unsigned SrcNumElts =
3701 unsigned Scale = SrcNumElts / DstNumElts;
3708 for (
unsigned i = 0; i != DstNumElts; ++i) {
3709 Idxs[i] = i + (Imm * DstNumElts);
3711 Rep = Builder.CreateShuffleVector(Op0, Op0, Idxs);
3717 }
else if (Name.starts_with(
"avx512.mask.perm.df.") ||
3718 Name.starts_with(
"avx512.mask.perm.di.")) {
3722 unsigned NumElts = VecTy->getNumElements();
3725 for (
unsigned i = 0; i != NumElts; ++i)
3726 Idxs[i] = (i & ~0x3) + ((Imm >> (2 * (i & 0x3))) & 3);
3728 Rep = Builder.CreateShuffleVector(Op0, Op0, Idxs);
3733 }
else if (Name.starts_with(
"avx.vperm2f128.") || Name ==
"avx2.vperm2i128") {
3745 unsigned HalfSize = NumElts / 2;
3757 unsigned StartIndex = (Imm & 0x01) ? HalfSize : 0;
3758 for (
unsigned i = 0; i < HalfSize; ++i)
3759 ShuffleMask[i] = StartIndex + i;
3762 StartIndex = (Imm & 0x10) ? HalfSize : 0;
3763 for (
unsigned i = 0; i < HalfSize; ++i)
3764 ShuffleMask[i + HalfSize] = NumElts + StartIndex + i;
3766 Rep = Builder.CreateShuffleVector(V0,
V1, ShuffleMask);
3768 }
else if (Name.starts_with(
"avx.vpermil.") || Name ==
"sse2.pshuf.d" ||
3769 Name.starts_with(
"avx512.mask.vpermil.p") ||
3770 Name.starts_with(
"avx512.mask.pshuf.d.")) {
3774 unsigned NumElts = VecTy->getNumElements();
3776 unsigned IdxSize = 64 / VecTy->getScalarSizeInBits();
3777 unsigned IdxMask = ((1 << IdxSize) - 1);
3783 for (
unsigned i = 0; i != NumElts; ++i)
3784 Idxs[i] = ((Imm >> ((i * IdxSize) % 8)) & IdxMask) | (i & ~IdxMask);
3786 Rep = Builder.CreateShuffleVector(Op0, Op0, Idxs);
3791 }
else if (Name ==
"sse2.pshufl.w" ||
3792 Name.starts_with(
"avx512.mask.pshufl.w.")) {
3797 if (Name ==
"sse2.pshufl.w" && NumElts % 8 != 0)
3801 for (
unsigned l = 0; l != NumElts; l += 8) {
3802 for (
unsigned i = 0; i != 4; ++i)
3803 Idxs[i + l] = ((Imm >> (2 * i)) & 0x3) + l;
3804 for (
unsigned i = 4; i != 8; ++i)
3805 Idxs[i + l] = i + l;
3808 Rep = Builder.CreateShuffleVector(Op0, Op0, Idxs);
3813 }
else if (Name ==
"sse2.pshufh.w" ||
3814 Name.starts_with(
"avx512.mask.pshufh.w.")) {
3819 if (Name ==
"sse2.pshufh.w" && NumElts % 8 != 0)
3823 for (
unsigned l = 0; l != NumElts; l += 8) {
3824 for (
unsigned i = 0; i != 4; ++i)
3825 Idxs[i + l] = i + l;
3826 for (
unsigned i = 0; i != 4; ++i)
3827 Idxs[i + l + 4] = ((Imm >> (2 * i)) & 0x3) + 4 + l;
3830 Rep = Builder.CreateShuffleVector(Op0, Op0, Idxs);
3835 }
else if (Name.starts_with(
"avx512.mask.shuf.p")) {
3842 unsigned HalfLaneElts = NumLaneElts / 2;
3845 for (
unsigned i = 0; i != NumElts; ++i) {
3847 Idxs[i] = i - (i % NumLaneElts);
3849 if ((i % NumLaneElts) >= HalfLaneElts)
3853 Idxs[i] += (Imm >> ((i * HalfLaneElts) % 8)) & ((1 << HalfLaneElts) - 1);
3856 Rep = Builder.CreateShuffleVector(Op0, Op1, Idxs);
3860 }
else if (Name.starts_with(
"avx512.mask.movddup") ||
3861 Name.starts_with(
"avx512.mask.movshdup") ||
3862 Name.starts_with(
"avx512.mask.movsldup")) {
3868 if (Name.starts_with(
"avx512.mask.movshdup."))
3872 for (
unsigned l = 0; l != NumElts; l += NumLaneElts)
3873 for (
unsigned i = 0; i != NumLaneElts; i += 2) {
3874 Idxs[i + l + 0] = i + l +
Offset;
3875 Idxs[i + l + 1] = i + l +
Offset;
3878 Rep = Builder.CreateShuffleVector(Op0, Op0, Idxs);
3882 }
else if (Name.starts_with(
"avx512.mask.punpckl") ||
3883 Name.starts_with(
"avx512.mask.unpckl.")) {
3890 for (
int l = 0; l != NumElts; l += NumLaneElts)
3891 for (
int i = 0; i != NumLaneElts; ++i)
3892 Idxs[i + l] = l + (i / 2) + NumElts * (i % 2);
3894 Rep = Builder.CreateShuffleVector(Op0, Op1, Idxs);
3898 }
else if (Name.starts_with(
"avx512.mask.punpckh") ||
3899 Name.starts_with(
"avx512.mask.unpckh.")) {
3906 for (
int l = 0; l != NumElts; l += NumLaneElts)
3907 for (
int i = 0; i != NumLaneElts; ++i)
3908 Idxs[i + l] = (NumLaneElts / 2) + l + (i / 2) + NumElts * (i % 2);
3910 Rep = Builder.CreateShuffleVector(Op0, Op1, Idxs);
3914 }
else if (Name.starts_with(
"avx512.mask.and.") ||
3915 Name.starts_with(
"avx512.mask.pand.")) {
3918 Rep = Builder.CreateAnd(Builder.CreateBitCast(CI->
getArgOperand(0), ITy),
3920 Rep = Builder.CreateBitCast(Rep, FTy);
3923 }
else if (Name.starts_with(
"avx512.mask.andn.") ||
3924 Name.starts_with(
"avx512.mask.pandn.")) {
3927 Rep = Builder.CreateNot(Builder.CreateBitCast(CI->
getArgOperand(0), ITy));
3928 Rep = Builder.CreateAnd(Rep,
3930 Rep = Builder.CreateBitCast(Rep, FTy);
3933 }
else if (Name.starts_with(
"avx512.mask.or.") ||
3934 Name.starts_with(
"avx512.mask.por.")) {
3937 Rep = Builder.CreateOr(Builder.CreateBitCast(CI->
getArgOperand(0), ITy),
3939 Rep = Builder.CreateBitCast(Rep, FTy);
3942 }
else if (Name.starts_with(
"avx512.mask.xor.") ||
3943 Name.starts_with(
"avx512.mask.pxor.")) {
3946 Rep = Builder.CreateXor(Builder.CreateBitCast(CI->
getArgOperand(0), ITy),
3948 Rep = Builder.CreateBitCast(Rep, FTy);
3951 }
else if (Name.starts_with(
"avx512.mask.padd.")) {
3955 }
else if (Name.starts_with(
"avx512.mask.psub.")) {
3959 }
else if (Name.starts_with(
"avx512.mask.pmull.")) {
3963 }
else if (Name.starts_with(
"avx512.mask.add.p")) {
3964 if (Name.ends_with(
".512")) {
3966 if (Name[17] ==
's')
3967 IID = Intrinsic::x86_avx512_add_ps_512;
3969 IID = Intrinsic::x86_avx512_add_pd_512;
3971 Rep = Builder.CreateIntrinsic(
3979 }
else if (Name.starts_with(
"avx512.mask.div.p")) {
3980 if (Name.ends_with(
".512")) {
3982 if (Name[17] ==
's')
3983 IID = Intrinsic::x86_avx512_div_ps_512;
3985 IID = Intrinsic::x86_avx512_div_pd_512;
3987 Rep = Builder.CreateIntrinsic(
3995 }
else if (Name.starts_with(
"avx512.mask.mul.p")) {
3996 if (Name.ends_with(
".512")) {
3998 if (Name[17] ==
's')
3999 IID = Intrinsic::x86_avx512_mul_ps_512;
4001 IID = Intrinsic::x86_avx512_mul_pd_512;
4003 Rep = Builder.CreateIntrinsic(
4011 }
else if (Name.starts_with(
"avx512.mask.sub.p")) {
4012 if (Name.ends_with(
".512")) {
4014 if (Name[17] ==
's')
4015 IID = Intrinsic::x86_avx512_sub_ps_512;
4017 IID = Intrinsic::x86_avx512_sub_pd_512;
4019 Rep = Builder.CreateIntrinsic(
4027 }
else if ((Name.starts_with(
"avx512.mask.max.p") ||
4028 Name.starts_with(
"avx512.mask.min.p")) &&
4029 Name.drop_front(18) ==
".512") {
4030 bool IsDouble = Name[17] ==
'd';
4031 bool IsMin = Name[13] ==
'i';
4033 {Intrinsic::x86_avx512_max_ps_512, Intrinsic::x86_avx512_max_pd_512},
4034 {Intrinsic::x86_avx512_min_ps_512, Intrinsic::x86_avx512_min_pd_512}};
4037 Rep = Builder.CreateIntrinsic(
4042 }
else if (Name.starts_with(
"avx512.mask.lzcnt.")) {
4044 Builder.CreateIntrinsic(Intrinsic::ctlz, CI->
getType(),
4045 {CI->getArgOperand(0), Builder.getInt1(false)});
4048 }
else if (Name.starts_with(
"avx512.mask.psll")) {
4049 bool IsImmediate = Name[16] ==
'i' || (Name.size() > 18 && Name[18] ==
'i');
4050 bool IsVariable = Name[16] ==
'v';
4051 char Size = Name[16] ==
'.' ? Name[17]
4052 : Name[17] ==
'.' ? Name[18]
4053 : Name[18] ==
'.' ? Name[19]
4057 if (IsVariable && Name[17] !=
'.') {
4058 if (
Size ==
'd' && Name[17] ==
'2')
4059 IID = Intrinsic::x86_avx2_psllv_q;
4060 else if (
Size ==
'd' && Name[17] ==
'4')
4061 IID = Intrinsic::x86_avx2_psllv_q_256;
4062 else if (
Size ==
's' && Name[17] ==
'4')
4063 IID = Intrinsic::x86_avx2_psllv_d;
4064 else if (
Size ==
's' && Name[17] ==
'8')
4065 IID = Intrinsic::x86_avx2_psllv_d_256;
4066 else if (
Size ==
'h' && Name[17] ==
'8')
4067 IID = Intrinsic::x86_avx512_psllv_w_128;
4068 else if (
Size ==
'h' && Name[17] ==
'1')
4069 IID = Intrinsic::x86_avx512_psllv_w_256;
4070 else if (Name[17] ==
'3' && Name[18] ==
'2')
4071 IID = Intrinsic::x86_avx512_psllv_w_512;
4074 }
else if (Name.ends_with(
".128")) {
4076 IID = IsImmediate ? Intrinsic::x86_sse2_pslli_d
4077 : Intrinsic::x86_sse2_psll_d;
4078 else if (
Size ==
'q')
4079 IID = IsImmediate ? Intrinsic::x86_sse2_pslli_q
4080 : Intrinsic::x86_sse2_psll_q;
4081 else if (
Size ==
'w')
4082 IID = IsImmediate ? Intrinsic::x86_sse2_pslli_w
4083 : Intrinsic::x86_sse2_psll_w;
4086 }
else if (Name.ends_with(
".256")) {
4088 IID = IsImmediate ? Intrinsic::x86_avx2_pslli_d
4089 : Intrinsic::x86_avx2_psll_d;
4090 else if (
Size ==
'q')
4091 IID = IsImmediate ? Intrinsic::x86_avx2_pslli_q
4092 : Intrinsic::x86_avx2_psll_q;
4093 else if (
Size ==
'w')
4094 IID = IsImmediate ? Intrinsic::x86_avx2_pslli_w
4095 : Intrinsic::x86_avx2_psll_w;
4100 IID = IsImmediate ? Intrinsic::x86_avx512_pslli_d_512
4101 : IsVariable ? Intrinsic::x86_avx512_psllv_d_512
4102 : Intrinsic::x86_avx512_psll_d_512;
4103 else if (
Size ==
'q')
4104 IID = IsImmediate ? Intrinsic::x86_avx512_pslli_q_512
4105 : IsVariable ? Intrinsic::x86_avx512_psllv_q_512
4106 : Intrinsic::x86_avx512_psll_q_512;
4107 else if (
Size ==
'w')
4108 IID = IsImmediate ? Intrinsic::x86_avx512_pslli_w_512
4109 : Intrinsic::x86_avx512_psll_w_512;
4115 }
else if (Name.starts_with(
"avx512.mask.psrl")) {
4116 bool IsImmediate = Name[16] ==
'i' || (Name.size() > 18 && Name[18] ==
'i');
4117 bool IsVariable = Name[16] ==
'v';
4118 char Size = Name[16] ==
'.' ? Name[17]
4119 : Name[17] ==
'.' ? Name[18]
4120 : Name[18] ==
'.' ? Name[19]
4124 if (IsVariable && Name[17] !=
'.') {
4125 if (
Size ==
'd' && Name[17] ==
'2')
4126 IID = Intrinsic::x86_avx2_psrlv_q;
4127 else if (
Size ==
'd' && Name[17] ==
'4')
4128 IID = Intrinsic::x86_avx2_psrlv_q_256;
4129 else if (
Size ==
's' && Name[17] ==
'4')
4130 IID = Intrinsic::x86_avx2_psrlv_d;
4131 else if (
Size ==
's' && Name[17] ==
'8')
4132 IID = Intrinsic::x86_avx2_psrlv_d_256;
4133 else if (
Size ==
'h' && Name[17] ==
'8')
4134 IID = Intrinsic::x86_avx512_psrlv_w_128;
4135 else if (
Size ==
'h' && Name[17] ==
'1')
4136 IID = Intrinsic::x86_avx512_psrlv_w_256;
4137 else if (Name[17] ==
'3' && Name[18] ==
'2')
4138 IID = Intrinsic::x86_avx512_psrlv_w_512;
4141 }
else if (Name.ends_with(
".128")) {
4143 IID = IsImmediate ? Intrinsic::x86_sse2_psrli_d
4144 : Intrinsic::x86_sse2_psrl_d;
4145 else if (
Size ==
'q')
4146 IID = IsImmediate ? Intrinsic::x86_sse2_psrli_q
4147 : Intrinsic::x86_sse2_psrl_q;
4148 else if (
Size ==
'w')
4149 IID = IsImmediate ? Intrinsic::x86_sse2_psrli_w
4150 : Intrinsic::x86_sse2_psrl_w;
4153 }
else if (Name.ends_with(
".256")) {
4155 IID = IsImmediate ? Intrinsic::x86_avx2_psrli_d
4156 : Intrinsic::x86_avx2_psrl_d;
4157 else if (
Size ==
'q')
4158 IID = IsImmediate ? Intrinsic::x86_avx2_psrli_q
4159 : Intrinsic::x86_avx2_psrl_q;
4160 else if (
Size ==
'w')
4161 IID = IsImmediate ? Intrinsic::x86_avx2_psrli_w
4162 : Intrinsic::x86_avx2_psrl_w;
4167 IID = IsImmediate ? Intrinsic::x86_avx512_psrli_d_512
4168 : IsVariable ? Intrinsic::x86_avx512_psrlv_d_512
4169 : Intrinsic::x86_avx512_psrl_d_512;
4170 else if (
Size ==
'q')
4171 IID = IsImmediate ? Intrinsic::x86_avx512_psrli_q_512
4172 : IsVariable ? Intrinsic::x86_avx512_psrlv_q_512
4173 : Intrinsic::x86_avx512_psrl_q_512;
4174 else if (
Size ==
'w')
4175 IID = IsImmediate ? Intrinsic::x86_avx512_psrli_w_512
4176 : Intrinsic::x86_avx512_psrl_w_512;
4182 }
else if (Name.starts_with(
"avx512.mask.psra")) {
4183 bool IsImmediate = Name[16] ==
'i' || (Name.size() > 18 && Name[18] ==
'i');
4184 bool IsVariable = Name[16] ==
'v';
4185 char Size = Name[16] ==
'.' ? Name[17]
4186 : Name[17] ==
'.' ? Name[18]
4187 : Name[18] ==
'.' ? Name[19]
4191 if (IsVariable && Name[17] !=
'.') {
4192 if (
Size ==
's' && Name[17] ==
'4')
4193 IID = Intrinsic::x86_avx2_psrav_d;
4194 else if (
Size ==
's' && Name[17] ==
'8')
4195 IID = Intrinsic::x86_avx2_psrav_d_256;
4196 else if (
Size ==
'h' && Name[17] ==
'8')
4197 IID = Intrinsic::x86_avx512_psrav_w_128;
4198 else if (
Size ==
'h' && Name[17] ==
'1')
4199 IID = Intrinsic::x86_avx512_psrav_w_256;
4200 else if (Name[17] ==
'3' && Name[18] ==
'2')
4201 IID = Intrinsic::x86_avx512_psrav_w_512;
4204 }
else if (Name.ends_with(
".128")) {
4206 IID = IsImmediate ? Intrinsic::x86_sse2_psrai_d
4207 : Intrinsic::x86_sse2_psra_d;
4208 else if (
Size ==
'q')
4209 IID = IsImmediate ? Intrinsic::x86_avx512_psrai_q_128
4210 : IsVariable ? Intrinsic::x86_avx512_psrav_q_128
4211 : Intrinsic::x86_avx512_psra_q_128;
4212 else if (
Size ==
'w')
4213 IID = IsImmediate ? Intrinsic::x86_sse2_psrai_w
4214 : Intrinsic::x86_sse2_psra_w;
4217 }
else if (Name.ends_with(
".256")) {
4219 IID = IsImmediate ? Intrinsic::x86_avx2_psrai_d
4220 : Intrinsic::x86_avx2_psra_d;
4221 else if (
Size ==
'q')
4222 IID = IsImmediate ? Intrinsic::x86_avx512_psrai_q_256
4223 : IsVariable ? Intrinsic::x86_avx512_psrav_q_256
4224 : Intrinsic::x86_avx512_psra_q_256;
4225 else if (
Size ==
'w')
4226 IID = IsImmediate ? Intrinsic::x86_avx2_psrai_w
4227 : Intrinsic::x86_avx2_psra_w;
4232 IID = IsImmediate ? Intrinsic::x86_avx512_psrai_d_512
4233 : IsVariable ? Intrinsic::x86_avx512_psrav_d_512
4234 : Intrinsic::x86_avx512_psra_d_512;
4235 else if (
Size ==
'q')
4236 IID = IsImmediate ? Intrinsic::x86_avx512_psrai_q_512
4237 : IsVariable ? Intrinsic::x86_avx512_psrav_q_512
4238 : Intrinsic::x86_avx512_psra_q_512;
4239 else if (
Size ==
'w')
4240 IID = IsImmediate ? Intrinsic::x86_avx512_psrai_w_512
4241 : Intrinsic::x86_avx512_psra_w_512;
4247 }
else if (Name.starts_with(
"avx512.mask.move.s")) {
4249 }
else if (Name.starts_with(
"avx512.cvtmask2")) {
4251 }
else if (Name.ends_with(
".movntdqa")) {
4255 LoadInst *LI = Builder.CreateAlignedLoad(
4260 }
else if (Name.starts_with(
"fma.vfmadd.") ||
4261 Name.starts_with(
"fma.vfmsub.") ||
4262 Name.starts_with(
"fma.vfnmadd.") ||
4263 Name.starts_with(
"fma.vfnmsub.")) {
4264 bool NegMul = Name[6] ==
'n';
4265 bool NegAcc = NegMul ? Name[8] ==
's' : Name[7] ==
's';
4266 bool IsScalar = NegMul ? Name[12] ==
's' : Name[11] ==
's';
4277 if (NegMul && !IsScalar)
4278 Ops[0] = Builder.CreateFNeg(
Ops[0]);
4279 if (NegMul && IsScalar)
4280 Ops[1] = Builder.CreateFNeg(
Ops[1]);
4282 Ops[2] = Builder.CreateFNeg(
Ops[2]);
4284 Rep = Builder.CreateIntrinsic(Intrinsic::fma,
Ops[0]->
getType(),
Ops);
4288 }
else if (Name.starts_with(
"fma4.vfmadd.s")) {
4296 Rep = Builder.CreateIntrinsic(Intrinsic::fma,
Ops[0]->
getType(),
Ops);
4300 }
else if (Name.starts_with(
"avx512.mask.vfmadd.s") ||
4301 Name.starts_with(
"avx512.maskz.vfmadd.s") ||
4302 Name.starts_with(
"avx512.mask3.vfmadd.s") ||
4303 Name.starts_with(
"avx512.mask3.vfmsub.s") ||
4304 Name.starts_with(
"avx512.mask3.vfnmsub.s")) {
4305 bool IsMask3 = Name[11] ==
'3';
4306 bool IsMaskZ = Name[11] ==
'z';
4308 Name = Name.drop_front(IsMask3 || IsMaskZ ? 13 : 12);
4309 bool NegMul = Name[2] ==
'n';
4310 bool NegAcc = NegMul ? Name[4] ==
's' : Name[3] ==
's';
4316 if (NegMul && (IsMask3 || IsMaskZ))
4317 A = Builder.CreateFNeg(
A);
4318 if (NegMul && !(IsMask3 || IsMaskZ))
4319 B = Builder.CreateFNeg(
B);
4321 C = Builder.CreateFNeg(
C);
4323 A = Builder.CreateExtractElement(
A, (
uint64_t)0);
4324 B = Builder.CreateExtractElement(
B, (
uint64_t)0);
4325 C = Builder.CreateExtractElement(
C, (
uint64_t)0);
4332 if (Name.back() ==
'd')
4333 IID = Intrinsic::x86_avx512_vfmadd_f64;
4335 IID = Intrinsic::x86_avx512_vfmadd_f32;
4336 Rep = Builder.CreateIntrinsic(IID,
Ops);
4338 Rep = Builder.CreateFMA(
A,
B,
C);
4347 if (NegAcc && IsMask3)
4352 Rep = Builder.CreateInsertElement(CI->
getArgOperand(IsMask3 ? 2 : 0), Rep,
4354 }
else if (Name.starts_with(
"avx512.mask.vfmadd.p") ||
4355 Name.starts_with(
"avx512.mask.vfnmadd.p") ||
4356 Name.starts_with(
"avx512.mask.vfnmsub.p") ||
4357 Name.starts_with(
"avx512.mask3.vfmadd.p") ||
4358 Name.starts_with(
"avx512.mask3.vfmsub.p") ||
4359 Name.starts_with(
"avx512.mask3.vfnmsub.p") ||
4360 Name.starts_with(
"avx512.maskz.vfmadd.p")) {
4361 bool IsMask3 = Name[11] ==
'3';
4362 bool IsMaskZ = Name[11] ==
'z';
4364 Name = Name.drop_front(IsMask3 || IsMaskZ ? 13 : 12);
4365 bool NegMul = Name[2] ==
'n';
4366 bool NegAcc = NegMul ? Name[4] ==
's' : Name[3] ==
's';
4372 if (NegMul && (IsMask3 || IsMaskZ))
4373 A = Builder.CreateFNeg(
A);
4374 if (NegMul && !(IsMask3 || IsMaskZ))
4375 B = Builder.CreateFNeg(
B);
4377 C = Builder.CreateFNeg(
C);
4384 if (Name[Name.size() - 5] ==
's')
4385 IID = Intrinsic::x86_avx512_vfmadd_ps_512;
4387 IID = Intrinsic::x86_avx512_vfmadd_pd_512;
4391 Rep = Builder.CreateFMA(
A,
B,
C);
4399 }
else if (Name.starts_with(
"fma.vfmsubadd.p")) {
4403 if (VecWidth == 128 && EltWidth == 32)
4404 IID = Intrinsic::x86_fma_vfmaddsub_ps;
4405 else if (VecWidth == 256 && EltWidth == 32)
4406 IID = Intrinsic::x86_fma_vfmaddsub_ps_256;
4407 else if (VecWidth == 128 && EltWidth == 64)
4408 IID = Intrinsic::x86_fma_vfmaddsub_pd;
4409 else if (VecWidth == 256 && EltWidth == 64)
4410 IID = Intrinsic::x86_fma_vfmaddsub_pd_256;
4416 Ops[2] = Builder.CreateFNeg(
Ops[2]);
4417 Rep = Builder.CreateIntrinsic(IID,
Ops);
4418 }
else if (Name.starts_with(
"avx512.mask.vfmaddsub.p") ||
4419 Name.starts_with(
"avx512.mask3.vfmaddsub.p") ||
4420 Name.starts_with(
"avx512.maskz.vfmaddsub.p") ||
4421 Name.starts_with(
"avx512.mask3.vfmsubadd.p")) {
4422 bool IsMask3 = Name[11] ==
'3';
4423 bool IsMaskZ = Name[11] ==
'z';
4425 Name = Name.drop_front(IsMask3 || IsMaskZ ? 13 : 12);
4426 bool IsSubAdd = Name[3] ==
's';
4430 if (Name[Name.size() - 5] ==
's')
4431 IID = Intrinsic::x86_avx512_vfmaddsub_ps_512;
4433 IID = Intrinsic::x86_avx512_vfmaddsub_pd_512;
4438 Ops[2] = Builder.CreateFNeg(
Ops[2]);
4440 Rep = Builder.CreateIntrinsic(IID,
Ops);
4449 Value *Odd = Builder.CreateCall(FMA,
Ops);
4450 Ops[2] = Builder.CreateFNeg(
Ops[2]);
4451 Value *Even = Builder.CreateCall(FMA,
Ops);
4457 for (
int i = 0; i != NumElts; ++i)
4458 Idxs[i] = i + (i % 2) * NumElts;
4460 Rep = Builder.CreateShuffleVector(Even, Odd, Idxs);
4468 }
else if (Name.starts_with(
"avx512.mask.pternlog.") ||
4469 Name.starts_with(
"avx512.maskz.pternlog.")) {
4470 bool ZeroMask = Name[11] ==
'z';
4474 if (VecWidth == 128 && EltWidth == 32)
4475 IID = Intrinsic::x86_avx512_pternlog_d_128;
4476 else if (VecWidth == 256 && EltWidth == 32)
4477 IID = Intrinsic::x86_avx512_pternlog_d_256;
4478 else if (VecWidth == 512 && EltWidth == 32)
4479 IID = Intrinsic::x86_avx512_pternlog_d_512;
4480 else if (VecWidth == 128 && EltWidth == 64)
4481 IID = Intrinsic::x86_avx512_pternlog_q_128;
4482 else if (VecWidth == 256 && EltWidth == 64)
4483 IID = Intrinsic::x86_avx512_pternlog_q_256;
4484 else if (VecWidth == 512 && EltWidth == 64)
4485 IID = Intrinsic::x86_avx512_pternlog_q_512;
4491 Rep = Builder.CreateIntrinsic(IID, Args);
4495 }
else if (Name.starts_with(
"avx512.mask.vpmadd52") ||
4496 Name.starts_with(
"avx512.maskz.vpmadd52")) {
4497 bool ZeroMask = Name[11] ==
'z';
4498 bool High = Name[20] ==
'h' || Name[21] ==
'h';
4501 if (VecWidth == 128 && !
High)
4502 IID = Intrinsic::x86_avx512_vpmadd52l_uq_128;
4503 else if (VecWidth == 256 && !
High)
4504 IID = Intrinsic::x86_avx512_vpmadd52l_uq_256;
4505 else if (VecWidth == 512 && !
High)
4506 IID = Intrinsic::x86_avx512_vpmadd52l_uq_512;
4507 else if (VecWidth == 128 &&
High)
4508 IID = Intrinsic::x86_avx512_vpmadd52h_uq_128;
4509 else if (VecWidth == 256 &&
High)
4510 IID = Intrinsic::x86_avx512_vpmadd52h_uq_256;
4511 else if (VecWidth == 512 &&
High)
4512 IID = Intrinsic::x86_avx512_vpmadd52h_uq_512;
4518 Rep = Builder.CreateIntrinsic(IID, Args);
4522 }
else if (Name.starts_with(
"avx512.mask.vpermi2var.") ||
4523 Name.starts_with(
"avx512.mask.vpermt2var.") ||
4524 Name.starts_with(
"avx512.maskz.vpermt2var.")) {
4525 bool ZeroMask = Name[11] ==
'z';
4526 bool IndexForm = Name[17] ==
'i';
4528 }
else if (Name.starts_with(
"avx512.mask.vpdpbusd.") ||
4529 Name.starts_with(
"avx512.maskz.vpdpbusd.") ||
4530 Name.starts_with(
"avx512.mask.vpdpbusds.") ||
4531 Name.starts_with(
"avx512.maskz.vpdpbusds.")) {
4532 bool ZeroMask = Name[11] ==
'z';
4533 bool IsSaturating = Name[ZeroMask ? 21 : 20] ==
's';
4536 if (VecWidth == 128 && !IsSaturating)
4537 IID = Intrinsic::x86_avx512_vpdpbusd_128;
4538 else if (VecWidth == 256 && !IsSaturating)
4539 IID = Intrinsic::x86_avx512_vpdpbusd_256;
4540 else if (VecWidth == 512 && !IsSaturating)
4541 IID = Intrinsic::x86_avx512_vpdpbusd_512;
4542 else if (VecWidth == 128 && IsSaturating)
4543 IID = Intrinsic::x86_avx512_vpdpbusds_128;
4544 else if (VecWidth == 256 && IsSaturating)
4545 IID = Intrinsic::x86_avx512_vpdpbusds_256;
4546 else if (VecWidth == 512 && IsSaturating)
4547 IID = Intrinsic::x86_avx512_vpdpbusds_512;
4557 if (Args[1]->
getType()->isVectorTy() &&
4560 ->isIntegerTy(32) &&
4561 Args[2]->
getType()->isVectorTy() &&
4564 ->isIntegerTy(32)) {
4565 Type *NewArgType =
nullptr;
4566 if (VecWidth == 128)
4568 else if (VecWidth == 256)
4570 else if (VecWidth == 512)
4576 Args[1] = Builder.CreateBitCast(Args[1], NewArgType);
4577 Args[2] = Builder.CreateBitCast(Args[2], NewArgType);
4580 Rep = Builder.CreateIntrinsic(IID, Args);
4584 }
else if (Name.starts_with(
"avx512.mask.vpdpwssd.") ||
4585 Name.starts_with(
"avx512.maskz.vpdpwssd.") ||
4586 Name.starts_with(
"avx512.mask.vpdpwssds.") ||
4587 Name.starts_with(
"avx512.maskz.vpdpwssds.")) {
4588 bool ZeroMask = Name[11] ==
'z';
4589 bool IsSaturating = Name[ZeroMask ? 21 : 20] ==
's';
4592 if (VecWidth == 128 && !IsSaturating)
4593 IID = Intrinsic::x86_avx512_vpdpwssd_128;
4594 else if (VecWidth == 256 && !IsSaturating)
4595 IID = Intrinsic::x86_avx512_vpdpwssd_256;
4596 else if (VecWidth == 512 && !IsSaturating)
4597 IID = Intrinsic::x86_avx512_vpdpwssd_512;
4598 else if (VecWidth == 128 && IsSaturating)
4599 IID = Intrinsic::x86_avx512_vpdpwssds_128;
4600 else if (VecWidth == 256 && IsSaturating)
4601 IID = Intrinsic::x86_avx512_vpdpwssds_256;
4602 else if (VecWidth == 512 && IsSaturating)
4603 IID = Intrinsic::x86_avx512_vpdpwssds_512;
4613 if (Args[1]->
getType()->isVectorTy() &&
4616 ->isIntegerTy(32) &&
4617 Args[2]->
getType()->isVectorTy() &&
4620 ->isIntegerTy(32)) {
4621 Type *NewArgType =
nullptr;
4622 if (VecWidth == 128)
4624 else if (VecWidth == 256)
4626 else if (VecWidth == 512)
4632 Args[1] = Builder.CreateBitCast(Args[1], NewArgType);
4633 Args[2] = Builder.CreateBitCast(Args[2], NewArgType);
4636 Rep = Builder.CreateIntrinsic(IID, Args);
4640 }
else if (Name ==
"addcarryx.u32" || Name ==
"addcarryx.u64" ||
4641 Name ==
"addcarry.u32" || Name ==
"addcarry.u64" ||
4642 Name ==
"subborrow.u32" || Name ==
"subborrow.u64") {
4644 if (Name[0] ==
'a' && Name.back() ==
'2')
4645 IID = Intrinsic::x86_addcarry_32;
4646 else if (Name[0] ==
'a' && Name.back() ==
'4')
4647 IID = Intrinsic::x86_addcarry_64;
4648 else if (Name[0] ==
's' && Name.back() ==
'2')
4649 IID = Intrinsic::x86_subborrow_32;
4650 else if (Name[0] ==
's' && Name.back() ==
'4')
4651 IID = Intrinsic::x86_subborrow_64;
4658 Value *NewCall = Builder.CreateIntrinsic(IID, Args);
4661 Value *
Data = Builder.CreateExtractValue(NewCall, 1);
4664 Value *CF = Builder.CreateExtractValue(NewCall, 0);
4668 }
else if (Name.starts_with(
"avx512.mask.") &&
4671 }
else if (Name.starts_with(
"bmi.pdep.")) {
4673 }
else if (Name.starts_with(
"bmi.pext.")) {
4683 if (Name.starts_with(
"neon.bfcvt")) {
4684 if (Name.starts_with(
"neon.bfcvtn2")) {
4686 std::iota(LoMask.
begin(), LoMask.
end(), 0);
4688 std::iota(ConcatMask.
begin(), ConcatMask.
end(), 0);
4689 Value *Inactive = Builder.CreateShuffleVector(CI->
getOperand(0), LoMask);
4692 return Builder.CreateShuffleVector(Inactive, Trunc, ConcatMask);
4693 }
else if (Name.starts_with(
"neon.bfcvtn")) {
4695 std::iota(ConcatMask.
begin(), ConcatMask.
end(), 0);
4699 dbgs() <<
"Trunc: " << *Trunc <<
"\n";
4700 return Builder.CreateShuffleVector(
4703 return Builder.CreateFPTrunc(CI->
getOperand(0),
4706 }
else if (Name.starts_with(
"sve.fcvt")) {
4709 .
Case(
"sve.fcvt.bf16f32", Intrinsic::aarch64_sve_fcvt_bf16f32_v2)
4710 .
Case(
"sve.fcvtnt.bf16f32",
4711 Intrinsic::aarch64_sve_fcvtnt_bf16f32_v2)
4723 if (Args[1]->
getType() != BadPredTy)
4726 Args[1] = Builder.CreateIntrinsic(Intrinsic::aarch64_sve_convert_to_svbool,
4727 BadPredTy, Args[1]);
4728 Args[1] = Builder.CreateIntrinsic(
4729 Intrinsic::aarch64_sve_convert_from_svbool, GoodPredTy, Args[1]);
4731 return Builder.CreateIntrinsic(NewID, Args,
nullptr,
4735 if (Name ==
"neon.vcvtfp2hf")
4736 return Builder.CreateBitCast(
4737 Builder.CreateFPTrunc(
4741 if (Name ==
"neon.vcvthf2fp")
4742 return Builder.CreateFPExt(
4743 Builder.CreateBitCast(
4753 if (Name ==
"mve.vctp64.old") {
4756 Value *VCTP = Builder.CreateIntrinsic(Intrinsic::arm_mve_vctp64, {},
4759 Value *C1 = Builder.CreateIntrinsic(
4760 Intrinsic::arm_mve_pred_v2i,
4762 return Builder.CreateIntrinsic(
4763 Intrinsic::arm_mve_pred_i2v,
4765 }
else if (Name ==
"mve.mull.int.predicated.v2i64.v4i32.v4i1" ||
4766 Name ==
"mve.vqdmull.predicated.v2i64.v4i32.v4i1" ||
4767 Name ==
"mve.vldr.gather.base.predicated.v2i64.v2i64.v4i1" ||
4768 Name ==
"mve.vldr.gather.base.wb.predicated.v2i64.v2i64.v4i1" ||
4770 "mve.vldr.gather.offset.predicated.v2i64.p0i64.v2i64.v4i1" ||
4771 Name ==
"mve.vldr.gather.offset.predicated.v2i64.p0.v2i64.v4i1" ||
4772 Name ==
"mve.vstr.scatter.base.predicated.v2i64.v2i64.v4i1" ||
4773 Name ==
"mve.vstr.scatter.base.wb.predicated.v2i64.v2i64.v4i1" ||
4775 "mve.vstr.scatter.offset.predicated.p0i64.v2i64.v2i64.v4i1" ||
4776 Name ==
"mve.vstr.scatter.offset.predicated.p0.v2i64.v2i64.v4i1" ||
4777 Name ==
"cde.vcx1q.predicated.v2i64.v4i1" ||
4778 Name ==
"cde.vcx1qa.predicated.v2i64.v4i1" ||
4779 Name ==
"cde.vcx2q.predicated.v2i64.v4i1" ||
4780 Name ==
"cde.vcx2qa.predicated.v2i64.v4i1" ||
4781 Name ==
"cde.vcx3q.predicated.v2i64.v4i1" ||
4782 Name ==
"cde.vcx3qa.predicated.v2i64.v4i1") {
4783 std::vector<Type *> Tys;
4787 case Intrinsic::arm_mve_mull_int_predicated:
4788 case Intrinsic::arm_mve_vqdmull_predicated:
4789 case Intrinsic::arm_mve_vldr_gather_base_predicated:
4792 case Intrinsic::arm_mve_vldr_gather_base_wb_predicated:
4793 case Intrinsic::arm_mve_vstr_scatter_base_predicated:
4794 case Intrinsic::arm_mve_vstr_scatter_base_wb_predicated:
4798 case Intrinsic::arm_mve_vldr_gather_offset_predicated:
4802 case Intrinsic::arm_mve_vstr_scatter_offset_predicated:
4806 case Intrinsic::arm_cde_vcx1q_predicated:
4807 case Intrinsic::arm_cde_vcx1qa_predicated:
4808 case Intrinsic::arm_cde_vcx2q_predicated:
4809 case Intrinsic::arm_cde_vcx2qa_predicated:
4810 case Intrinsic::arm_cde_vcx3q_predicated:
4811 case Intrinsic::arm_cde_vcx3qa_predicated:
4818 std::vector<Value *>
Ops;
4820 Type *Ty =
Op->getType();
4821 if (Ty->getScalarSizeInBits() == 1) {
4822 Value *C1 = Builder.CreateIntrinsic(
4823 Intrinsic::arm_mve_pred_v2i,
4825 Op = Builder.CreateIntrinsic(Intrinsic::arm_mve_pred_i2v, {V2I1Ty}, C1);
4830 return Builder.CreateIntrinsic(
ID, Tys,
Ops,
nullptr,
4845 auto UpgradeLegacyWMMAIUIntrinsicCall =
4850 Args.push_back(Builder.getFalse());
4854 F->getParent(),
F->getIntrinsicID(), OverloadTys);
4861 auto *NewCall =
cast<CallInst>(Builder.CreateCall(NewDecl, Args, Bundles));
4866 NewCall->copyMetadata(*CI);
4870 if (
F->getIntrinsicID() == Intrinsic::amdgcn_wmma_i32_16x16x64_iu8) {
4871 assert(CI->
arg_size() == 7 &&
"Legacy int_amdgcn_wmma_i32_16x16x64_iu8 "
4872 "intrinsic should have 7 arguments");
4875 return UpgradeLegacyWMMAIUIntrinsicCall(
F, CI, Builder, {
T1, T2});
4877 if (
F->getIntrinsicID() == Intrinsic::amdgcn_swmmac_i32_16x16x128_iu8) {
4878 assert(CI->
arg_size() == 8 &&
"Legacy int_amdgcn_swmmac_i32_16x16x128_iu8 "
4879 "intrinsic should have 8 arguments");
4884 return UpgradeLegacyWMMAIUIntrinsicCall(
F, CI, Builder, {
T1, T2, T3, T4});
4887 switch (
F->getIntrinsicID()) {
4890 case Intrinsic::amdgcn_wmma_f32_16x16x4_f32:
4891 case Intrinsic::amdgcn_wmma_f32_16x16x32_bf16:
4892 case Intrinsic::amdgcn_wmma_f32_16x16x32_f16:
4893 case Intrinsic::amdgcn_wmma_f16_16x16x32_f16:
4894 case Intrinsic::amdgcn_wmma_bf16_16x16x32_bf16:
4895 case Intrinsic::amdgcn_wmma_bf16f32_16x16x32_bf16: {
4910 if (
F->getIntrinsicID() == Intrinsic::amdgcn_wmma_bf16f32_16x16x32_bf16)
4913 F->getParent(),
F->getIntrinsicID(), Overloads);
4918 auto *NewCall =
cast<CallInst>(Builder.CreateCall(NewDecl, Args, Bundles));
4923 NewCall->copyMetadata(*CI);
4924 NewCall->takeName(CI);
4946 if (NumOperands < 3)
4959 bool IsVolatile =
false;
4963 if (NumOperands > 3)
4968 if (NumOperands > 5) {
4970 IsVolatile = !VolatileArg || !VolatileArg->
isZero();
4984 if (VT->getElementType()->isIntegerTy(16)) {
4987 Val = Builder.CreateBitCast(Val, AsBF16);
4995 Builder.CreateAtomicRMW(RMWOp, Ptr, Val, std::nullopt, Order, SSID);
4997 unsigned AddrSpace = PtrTy->getAddressSpace();
5000 RMW->
setMetadata(
"amdgpu.no.fine.grained.memory", EmptyMD);
5002 RMW->
setMetadata(
"amdgpu.ignore.denormal.mode", EmptyMD);
5007 MDNode *RangeNotPrivate =
5010 RMW->
setMetadata(LLVMContext::MD_noalias_addrspace, RangeNotPrivate);
5016 return Builder.CreateBitCast(RMW, RetTy);
5037 return MAV->getMetadata();
5046 if (Name ==
"label") {
5048 }
else if (Name ==
"assign") {
5055 }
else if (Name ==
"declare") {
5059 }
else if (Name ==
"addr") {
5069 unwrapMAVOp(CI, 1), ExprNode,
nullptr,
nullptr,
nullptr);
5070 }
else if (Name ==
"value") {
5073 unsigned ExprOp = 2;
5088 assert(DR &&
"Unhandled intrinsic kind in upgrade to DbgRecord");
5096 int64_t OffsetVal =
Offset->getSExtValue();
5097 return Builder.CreateIntrinsic(OffsetVal >= 0
5098 ? Intrinsic::vector_splice_left
5099 : Intrinsic::vector_splice_right,
5101 {CI->getArgOperand(0), CI->getArgOperand(1),
5102 Builder.getInt32(std::abs(OffsetVal))});
5107 if (Name.starts_with(
"to.fp16")) {
5109 Builder.CreateFPTrunc(CI->
getArgOperand(0), Builder.getHalfTy());
5110 return Builder.CreateBitCast(Cast, CI->
getType());
5113 if (Name.starts_with(
"from.fp16")) {
5115 Builder.CreateBitCast(CI->
getArgOperand(0), Builder.getHalfTy());
5116 return Builder.CreateFPExt(Cast, CI->
getType());
5127 if (Defaults.empty())
5130 unsigned OldArgCount = CI->
arg_size();
5131 unsigned NewArgCount = NewFn->
arg_size();
5135 if (OldArgCount >= NewArgCount)
5143 if (OldArgCount < FirstDefault)
5148 for (
unsigned Idx = OldArgCount; Idx < NewArgCount; ++Idx) {
5149 assert(Idx >= FirstDefault && Idx - FirstDefault < Defaults.size() &&
5150 "missing argument outside the default range");
5151 Type *ParamTy = NewFT->getParamType(Idx);
5156 NewArgs.
push_back(ConstantInt::get(ParamTy, Defaults[Idx - FirstDefault]));
5162 CallInst *NewCall = Builder.CreateCall(NewFn, NewArgs, OpBundles);
5194 if (!Name.consume_front(
"llvm."))
5197 bool IsX86 = Name.consume_front(
"x86.");
5198 bool IsNVVM = Name.consume_front(
"nvvm.");
5199 bool IsAArch64 = Name.consume_front(
"aarch64.");
5200 bool IsARM = Name.consume_front(
"arm.");
5201 bool IsAMDGCN = Name.consume_front(
"amdgcn.");
5202 bool IsDbg = Name.consume_front(
"dbg.");
5204 (Name.consume_front(
"experimental.vector.splice") ||
5205 Name.consume_front(
"vector.splice")) &&
5206 !(Name.starts_with(
".left") || Name.starts_with(
".right"));
5207 Value *Rep =
nullptr;
5209 if (!IsX86 && Name ==
"stackprotectorcheck") {
5211 }
else if (IsNVVM) {
5215 }
else if (IsAArch64) {
5219 }
else if (IsAMDGCN) {
5223 }
else if (IsOldSplice) {
5225 }
else if (Name.consume_front(
"convert.")) {
5227 }
else if (Name ==
"lifetime.start.i64" || Name ==
"lifetime.end.i64") {
5240 const auto &DefaultCase = [&]() ->
void {
5248 "Unknown function for CallBase upgrade and isn't just a name change");
5256 "Return type must have changed");
5257 assert(OldST->getNumElements() ==
5259 "Must have same number of elements");
5262 CallInst *NewCI = Builder.CreateCall(NewFn, Args);
5265 for (
unsigned Idx = 0; Idx < OldST->getNumElements(); ++Idx) {
5266 Value *Elem = Builder.CreateExtractValue(NewCI, Idx);
5267 Res = Builder.CreateInsertValue(Res, Elem, Idx);
5291 case Intrinsic::arm_neon_vst1:
5292 case Intrinsic::arm_neon_vst2:
5293 case Intrinsic::arm_neon_vst3:
5294 case Intrinsic::arm_neon_vst4:
5295 case Intrinsic::arm_neon_vst2lane:
5296 case Intrinsic::arm_neon_vst3lane:
5297 case Intrinsic::arm_neon_vst4lane: {
5299 NewCall = Builder.CreateCall(NewFn, Args);
5302 case Intrinsic::aarch64_sve_bfmlalb_lane_v2:
5303 case Intrinsic::aarch64_sve_bfmlalt_lane_v2:
5304 case Intrinsic::aarch64_sve_bfdot_lane_v2: {
5309 NewCall = Builder.CreateCall(NewFn, Args);
5312 case Intrinsic::aarch64_sve_ld3_sret:
5313 case Intrinsic::aarch64_sve_ld4_sret:
5314 case Intrinsic::aarch64_sve_ld2_sret: {
5322 Name = Name.substr(5);
5329 unsigned MinElts = RetTy->getMinNumElements() /
N;
5331 Value *NewLdCall = Builder.CreateCall(NewFn, Args);
5333 for (
unsigned I = 0;
I <
N;
I++) {
5334 Value *SRet = Builder.CreateExtractValue(NewLdCall,
I);
5335 Ret = Builder.CreateInsertVector(RetTy, Ret, SRet,
I * MinElts);
5341 case Intrinsic::coro_end: {
5344 NewCall = Builder.CreateCall(NewFn, Args);
5348 case Intrinsic::vector_extract: {
5350 Name = Name.substr(5);
5351 if (!Name.starts_with(
"aarch64.sve.tuple.get")) {
5356 unsigned MinElts = RetTy->getMinNumElements();
5359 NewCall = Builder.CreateCall(NewFn, {CI->
getArgOperand(0), NewIdx});
5363 case Intrinsic::vector_insert: {
5365 Name = Name.substr(5);
5366 if (!Name.starts_with(
"aarch64.sve.tuple")) {
5370 if (Name.starts_with(
"aarch64.sve.tuple.set")) {
5375 NewCall = Builder.CreateCall(
5379 if (Name.starts_with(
"aarch64.sve.tuple.create")) {
5385 assert(
N > 1 &&
"Create is expected to be between 2-4");
5388 unsigned MinElts = RetTy->getMinNumElements() /
N;
5389 for (
unsigned I = 0;
I <
N;
I++) {
5391 Ret = Builder.CreateInsertVector(RetTy, Ret, V,
I * MinElts);
5398 case Intrinsic::arm_neon_bfdot:
5399 case Intrinsic::arm_neon_bfmmla:
5400 case Intrinsic::arm_neon_bfmlalb:
5401 case Intrinsic::arm_neon_bfmlalt:
5402 case Intrinsic::aarch64_neon_bfdot:
5403 case Intrinsic::aarch64_neon_bfmmla:
5404 case Intrinsic::aarch64_neon_bfmlalb:
5405 case Intrinsic::aarch64_neon_bfmlalt: {
5408 "Mismatch between function args and call args");
5409 size_t OperandWidth =
5411 assert((OperandWidth == 64 || OperandWidth == 128) &&
5412 "Unexpected operand width");
5414 auto Iter = CI->
args().begin();
5415 Args.push_back(*Iter++);
5416 Args.push_back(Builder.CreateBitCast(*Iter++, NewTy));
5417 Args.push_back(Builder.CreateBitCast(*Iter++, NewTy));
5418 NewCall = Builder.CreateCall(NewFn, Args);
5422 case Intrinsic::bitreverse:
5423 NewCall = Builder.CreateCall(NewFn, {CI->
getArgOperand(0)});
5426 case Intrinsic::ctlz:
5427 case Intrinsic::cttz: {
5434 Builder.CreateCall(NewFn, {CI->
getArgOperand(0), Builder.getFalse()});
5438 case Intrinsic::objectsize: {
5439 Value *NullIsUnknownSize =
5443 NewCall = Builder.CreateCall(
5448 case Intrinsic::ctpop:
5449 NewCall = Builder.CreateCall(NewFn, {CI->
getArgOperand(0)});
5451 case Intrinsic::dbg_value: {
5453 Name = Name.substr(5);
5455 if (Name.starts_with(
"dbg.addr")) {
5469 if (
Offset->isNullValue()) {
5470 NewCall = Builder.CreateCall(
5479 case Intrinsic::ptr_annotation:
5487 NewCall = Builder.CreateCall(
5496 case Intrinsic::var_annotation:
5503 NewCall = Builder.CreateCall(
5512 case Intrinsic::riscv_aes32dsi:
5513 case Intrinsic::riscv_aes32dsmi:
5514 case Intrinsic::riscv_aes32esi:
5515 case Intrinsic::riscv_aes32esmi:
5516 case Intrinsic::riscv_sm4ks:
5517 case Intrinsic::riscv_sm4ed: {
5527 Arg0 = Builder.CreateTrunc(Arg0, Builder.getInt32Ty());
5528 Arg1 = Builder.CreateTrunc(Arg1, Builder.getInt32Ty());
5534 NewCall = Builder.CreateCall(NewFn, {Arg0, Arg1, Arg2});
5535 Value *Res = NewCall;
5537 Res = Builder.CreateIntCast(NewCall, CI->
getType(),
true);
5543 case Intrinsic::nvvm_mapa_shared_cluster: {
5547 Value *Res = NewCall;
5548 Res = Builder.CreateAddrSpaceCast(
5555 case Intrinsic::nvvm_cp_async_bulk_global_to_shared_cluster:
5556 case Intrinsic::nvvm_cp_async_bulk_shared_cta_to_cluster: {
5559 Args[0] = Builder.CreateAddrSpaceCast(
5562 NewCall = Builder.CreateCall(NewFn, Args);
5568 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_im2col_3d:
5569 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_im2col_4d:
5570 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_im2col_5d:
5571 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_1d:
5572 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_2d:
5573 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_3d:
5574 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_4d:
5575 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_5d: {
5582 Args[0] = Builder.CreateAddrSpaceCast(
5591 Args.push_back(ConstantInt::get(Builder.getInt32Ty(), 0));
5593 NewCall = Builder.CreateCall(NewFn, Args);
5599 case Intrinsic::riscv_sha256sig0:
5600 case Intrinsic::riscv_sha256sig1:
5601 case Intrinsic::riscv_sha256sum0:
5602 case Intrinsic::riscv_sha256sum1:
5603 case Intrinsic::riscv_sm3p0:
5604 case Intrinsic::riscv_sm3p1: {
5611 Builder.CreateTrunc(CI->
getArgOperand(0), Builder.getInt32Ty());
5613 NewCall = Builder.CreateCall(NewFn, Arg);
5615 Builder.CreateIntCast(NewCall, CI->
getType(),
true);
5622 case Intrinsic::x86_xop_vfrcz_ss:
5623 case Intrinsic::x86_xop_vfrcz_sd:
5624 NewCall = Builder.CreateCall(NewFn, {CI->
getArgOperand(1)});
5627 case Intrinsic::x86_xop_vpermil2pd:
5628 case Intrinsic::x86_xop_vpermil2ps:
5629 case Intrinsic::x86_xop_vpermil2pd_256:
5630 case Intrinsic::x86_xop_vpermil2ps_256: {
5634 Args[2] = Builder.CreateBitCast(Args[2], IntIdxTy);
5635 NewCall = Builder.CreateCall(NewFn, Args);
5639 case Intrinsic::x86_sse41_ptestc:
5640 case Intrinsic::x86_sse41_ptestz:
5641 case Intrinsic::x86_sse41_ptestnzc: {
5655 Value *BC0 = Builder.CreateBitCast(Arg0, NewVecTy,
"cast");
5656 Value *BC1 = Builder.CreateBitCast(Arg1, NewVecTy,
"cast");
5658 NewCall = Builder.CreateCall(NewFn, {BC0, BC1});
5662 case Intrinsic::x86_rdtscp: {
5668 NewCall = Builder.CreateCall(NewFn);
5670 Value *
Data = Builder.CreateExtractValue(NewCall, 1);
5673 Value *TSC = Builder.CreateExtractValue(NewCall, 0);
5681 case Intrinsic::x86_sse41_insertps:
5682 case Intrinsic::x86_sse41_dppd:
5683 case Intrinsic::x86_sse41_dpps:
5684 case Intrinsic::x86_sse41_mpsadbw:
5685 case Intrinsic::x86_avx_dp_ps_256:
5686 case Intrinsic::x86_avx2_mpsadbw: {
5692 Args.back() = Builder.CreateTrunc(Args.back(),
Type::getInt8Ty(
C),
"trunc");
5693 NewCall = Builder.CreateCall(NewFn, Args);
5697 case Intrinsic::x86_avx512_mask_cmp_pd_128:
5698 case Intrinsic::x86_avx512_mask_cmp_pd_256:
5699 case Intrinsic::x86_avx512_mask_cmp_pd_512:
5700 case Intrinsic::x86_avx512_mask_cmp_ps_128:
5701 case Intrinsic::x86_avx512_mask_cmp_ps_256:
5702 case Intrinsic::x86_avx512_mask_cmp_ps_512: {
5708 NewCall = Builder.CreateCall(NewFn, Args);
5717 case Intrinsic::x86_avx512bf16_cvtne2ps2bf16_128:
5718 case Intrinsic::x86_avx512bf16_cvtne2ps2bf16_256:
5719 case Intrinsic::x86_avx512bf16_cvtne2ps2bf16_512:
5720 case Intrinsic::x86_avx512bf16_mask_cvtneps2bf16_128:
5721 case Intrinsic::x86_avx512bf16_cvtneps2bf16_256:
5722 case Intrinsic::x86_avx512bf16_cvtneps2bf16_512: {
5726 Intrinsic::x86_avx512bf16_mask_cvtneps2bf16_128)
5727 Args[1] = Builder.CreateBitCast(
5730 NewCall = Builder.CreateCall(NewFn, Args);
5731 Value *Res = Builder.CreateBitCast(
5739 case Intrinsic::x86_avx512bf16_dpbf16ps_128:
5740 case Intrinsic::x86_avx512bf16_dpbf16ps_256:
5741 case Intrinsic::x86_avx512bf16_dpbf16ps_512:{
5745 Args[1] = Builder.CreateBitCast(
5747 Args[2] = Builder.CreateBitCast(
5750 NewCall = Builder.CreateCall(NewFn, Args);
5754 case Intrinsic::thread_pointer: {
5755 NewCall = Builder.CreateCall(NewFn, {});
5759 case Intrinsic::memcpy:
5760 case Intrinsic::memmove:
5761 case Intrinsic::memset: {
5777 NewCall = Builder.CreateCall(NewFn, Args);
5779 AttributeList NewAttrs = AttributeList::get(
5780 C, OldAttrs.getFnAttrs(), OldAttrs.getRetAttrs(),
5781 {OldAttrs.getParamAttrs(0), OldAttrs.getParamAttrs(1),
5782 OldAttrs.getParamAttrs(2), OldAttrs.getParamAttrs(4)});
5787 MemCI->setDestAlignment(
Align->getMaybeAlignValue());
5790 MTI->setSourceAlignment(
Align->getMaybeAlignValue());
5794 case Intrinsic::masked_load:
5795 case Intrinsic::masked_gather:
5796 case Intrinsic::masked_store:
5797 case Intrinsic::masked_scatter: {
5803 auto GetMaybeAlign = [](
Value *
Op) {
5813 auto GetAlign = [&](
Value *
Op) {
5822 case Intrinsic::masked_load:
5823 NewCall = Builder.CreateMaskedLoad(
5827 case Intrinsic::masked_gather:
5828 NewCall = Builder.CreateMaskedGather(
5834 case Intrinsic::masked_store:
5835 NewCall = Builder.CreateMaskedStore(
5839 case Intrinsic::masked_scatter:
5840 NewCall = Builder.CreateMaskedScatter(
5842 DL.getValueOrABITypeAlignment(
5856 case Intrinsic::lifetime_start:
5857 case Intrinsic::lifetime_end: {
5869 NewCall = Builder.CreateLifetimeStart(Ptr);
5871 NewCall = Builder.CreateLifetimeEnd(Ptr);
5880 case Intrinsic::x86_avx512_vpdpbusd_128:
5881 case Intrinsic::x86_avx512_vpdpbusd_256:
5882 case Intrinsic::x86_avx512_vpdpbusd_512:
5883 case Intrinsic::x86_avx512_vpdpbusds_128:
5884 case Intrinsic::x86_avx512_vpdpbusds_256:
5885 case Intrinsic::x86_avx512_vpdpbusds_512:
5886 case Intrinsic::x86_avx2_vpdpbssd_128:
5887 case Intrinsic::x86_avx2_vpdpbssd_256:
5888 case Intrinsic::x86_avx10_vpdpbssd_512:
5889 case Intrinsic::x86_avx2_vpdpbssds_128:
5890 case Intrinsic::x86_avx2_vpdpbssds_256:
5891 case Intrinsic::x86_avx10_vpdpbssds_512:
5892 case Intrinsic::x86_avx2_vpdpbsud_128:
5893 case Intrinsic::x86_avx2_vpdpbsud_256:
5894 case Intrinsic::x86_avx10_vpdpbsud_512:
5895 case Intrinsic::x86_avx2_vpdpbsuds_128:
5896 case Intrinsic::x86_avx2_vpdpbsuds_256:
5897 case Intrinsic::x86_avx10_vpdpbsuds_512:
5898 case Intrinsic::x86_avx2_vpdpbuud_128:
5899 case Intrinsic::x86_avx2_vpdpbuud_256:
5900 case Intrinsic::x86_avx10_vpdpbuud_512:
5901 case Intrinsic::x86_avx2_vpdpbuuds_128:
5902 case Intrinsic::x86_avx2_vpdpbuuds_256:
5903 case Intrinsic::x86_avx10_vpdpbuuds_512: {
5908 Args[1] = Builder.CreateBitCast(Args[1], NewArgType);
5909 Args[2] = Builder.CreateBitCast(Args[2], NewArgType);
5911 NewCall = Builder.CreateCall(NewFn, Args);
5914 case Intrinsic::x86_avx512_vpdpwssd_128:
5915 case Intrinsic::x86_avx512_vpdpwssd_256:
5916 case Intrinsic::x86_avx512_vpdpwssd_512:
5917 case Intrinsic::x86_avx512_vpdpwssds_128:
5918 case Intrinsic::x86_avx512_vpdpwssds_256:
5919 case Intrinsic::x86_avx512_vpdpwssds_512:
5920 case Intrinsic::x86_avx2_vpdpwsud_128:
5921 case Intrinsic::x86_avx2_vpdpwsud_256:
5922 case Intrinsic::x86_avx10_vpdpwsud_512:
5923 case Intrinsic::x86_avx2_vpdpwsuds_128:
5924 case Intrinsic::x86_avx2_vpdpwsuds_256:
5925 case Intrinsic::x86_avx10_vpdpwsuds_512:
5926 case Intrinsic::x86_avx2_vpdpwusd_128:
5927 case Intrinsic::x86_avx2_vpdpwusd_256:
5928 case Intrinsic::x86_avx10_vpdpwusd_512:
5929 case Intrinsic::x86_avx2_vpdpwusds_128:
5930 case Intrinsic::x86_avx2_vpdpwusds_256:
5931 case Intrinsic::x86_avx10_vpdpwusds_512:
5932 case Intrinsic::x86_avx2_vpdpwuud_128:
5933 case Intrinsic::x86_avx2_vpdpwuud_256:
5934 case Intrinsic::x86_avx10_vpdpwuud_512:
5935 case Intrinsic::x86_avx2_vpdpwuuds_128:
5936 case Intrinsic::x86_avx2_vpdpwuuds_256:
5937 case Intrinsic::x86_avx10_vpdpwuuds_512:
5942 Args[1] = Builder.CreateBitCast(Args[1], NewArgType);
5943 Args[2] = Builder.CreateBitCast(Args[2], NewArgType);
5945 NewCall = Builder.CreateCall(NewFn, Args);
5948 assert(NewCall &&
"Should have either set this variable or returned through "
5949 "the default case");
5956 assert(
F &&
"Illegal attempt to upgrade a non-existent intrinsic.");
5970 F->eraseFromParent();
5976 if (NumOperands == 0)
5984 if (NumOperands == 3) {
5988 Metadata *Elts2[] = {ScalarType, ScalarType,
6002 if (
Opc != Instruction::BitCast)
6006 Type *SrcTy = V->getType();
6023 if (
Opc != Instruction::BitCast)
6026 Type *SrcTy =
C->getType();
6053 if (
NamedMDNode *ModFlags = M.getModuleFlagsMetadata()) {
6054 auto OpIt =
find_if(ModFlags->operands(), [](
const MDNode *Flag) {
6055 if (Flag->getNumOperands() < 3)
6057 if (MDString *K = dyn_cast_or_null<MDString>(Flag->getOperand(1)))
6058 return K->getString() ==
"Debug Info Version";
6061 if (OpIt != ModFlags->op_end()) {
6062 const MDOperand &ValOp = (*OpIt)->getOperand(2);
6069 bool BrokenDebugInfo =
false;
6072 if (!BrokenDebugInfo)
6078 M.getContext().diagnose(Diag);
6085 M.getContext().diagnose(DiagVersion);
6095 StringRef Vect3[3] = {DefaultValue, DefaultValue, DefaultValue};
6098 if (
F->hasFnAttribute(Attr)) {
6101 StringRef S =
F->getFnAttribute(Attr).getValueAsString();
6103 auto [Part, Rest] = S.
split(
',');
6109 const unsigned Dim = DimC -
'x';
6110 assert(Dim < 3 &&
"Unexpected dim char");
6120 F->addFnAttr(Attr, NewAttr);
6124 return S ==
"x" || S ==
"y" || S ==
"z";
6129 if (K ==
"kernel") {
6141 const unsigned Idx = (AlignIdxValuePair >> 16);
6142 const Align StackAlign =
Align(AlignIdxValuePair & 0xFFFF);
6147 if (K ==
"maxclusterrank" || K ==
"cluster_max_blocks") {
6152 if (K ==
"minctasm") {
6157 if (K ==
"maxnreg") {
6162 if (K.consume_front(
"maxntid") &&
isXYZ(K)) {
6166 if (K.consume_front(
"reqntid") &&
isXYZ(K)) {
6170 if (K.consume_front(
"cluster_dim_") &&
isXYZ(K)) {
6174 if (K ==
"grid_constant") {
6189 NamedMDNode *NamedMD = M.getNamedMetadata(
"nvvm.annotations");
6196 if (!SeenNodes.
insert(MD).second)
6203 assert((MD->getNumOperands() % 2) == 1 &&
"Invalid number of operands");
6210 for (
unsigned j = 1, je = MD->getNumOperands(); j < je; j += 2) {
6212 const MDOperand &V = MD->getOperand(j + 1);
6215 NewOperands.
append({K, V});
6218 if (NewOperands.
size() > 1)
6231 const char *MarkerKey =
"clang.arc.retainAutoreleasedReturnValueMarker";
6232 NamedMDNode *ModRetainReleaseMarker = M.getNamedMetadata(MarkerKey);
6233 if (ModRetainReleaseMarker) {
6239 ID->getString().split(ValueComp,
"#");
6240 if (ValueComp.
size() == 2) {
6241 std::string NewValue = ValueComp[0].str() +
";" + ValueComp[1].str();
6245 M.eraseNamedMetadata(ModRetainReleaseMarker);
6256 auto UpgradeToIntrinsic = [&](
const char *OldFunc,
6282 bool InvalidCast =
false;
6284 for (
unsigned I = 0, E = CI->
arg_size();
I != E; ++
I) {
6297 Arg = Builder.CreateBitCast(Arg, NewFuncTy->
getParamType(
I));
6299 Args.push_back(Arg);
6306 CallInst *NewCall = Builder.CreateCall(NewFuncTy, NewFn, Args);
6311 Value *NewRetVal = Builder.CreateBitCast(NewCall, CI->
getType());
6324 UpgradeToIntrinsic(
"clang.arc.use", llvm::Intrinsic::objc_clang_arc_use);
6332 std::pair<const char *, llvm::Intrinsic::ID> RuntimeFuncs[] = {
6333 {
"objc_autorelease", llvm::Intrinsic::objc_autorelease},
6334 {
"objc_autoreleasePoolPop", llvm::Intrinsic::objc_autoreleasePoolPop},
6335 {
"objc_autoreleasePoolPush", llvm::Intrinsic::objc_autoreleasePoolPush},
6336 {
"objc_autoreleaseReturnValue",
6337 llvm::Intrinsic::objc_autoreleaseReturnValue},
6338 {
"objc_copyWeak", llvm::Intrinsic::objc_copyWeak},
6339 {
"objc_destroyWeak", llvm::Intrinsic::objc_destroyWeak},
6340 {
"objc_initWeak", llvm::Intrinsic::objc_initWeak},
6341 {
"objc_loadWeak", llvm::Intrinsic::objc_loadWeak},
6342 {
"objc_loadWeakRetained", llvm::Intrinsic::objc_loadWeakRetained},
6343 {
"objc_moveWeak", llvm::Intrinsic::objc_moveWeak},
6344 {
"objc_release", llvm::Intrinsic::objc_release},
6345 {
"objc_retain", llvm::Intrinsic::objc_retain},
6346 {
"objc_retainAutorelease", llvm::Intrinsic::objc_retainAutorelease},
6347 {
"objc_retainAutoreleaseReturnValue",
6348 llvm::Intrinsic::objc_retainAutoreleaseReturnValue},
6349 {
"objc_retainAutoreleasedReturnValue",
6350 llvm::Intrinsic::objc_retainAutoreleasedReturnValue},
6351 {
"objc_retainBlock", llvm::Intrinsic::objc_retainBlock},
6352 {
"objc_storeStrong", llvm::Intrinsic::objc_storeStrong},
6353 {
"objc_storeWeak", llvm::Intrinsic::objc_storeWeak},
6354 {
"objc_unsafeClaimAutoreleasedReturnValue",
6355 llvm::Intrinsic::objc_unsafeClaimAutoreleasedReturnValue},
6356 {
"objc_retainedObject", llvm::Intrinsic::objc_retainedObject},
6357 {
"objc_unretainedObject", llvm::Intrinsic::objc_unretainedObject},
6358 {
"objc_unretainedPointer", llvm::Intrinsic::objc_unretainedPointer},
6359 {
"objc_retain_autorelease", llvm::Intrinsic::objc_retain_autorelease},
6360 {
"objc_sync_enter", llvm::Intrinsic::objc_sync_enter},
6361 {
"objc_sync_exit", llvm::Intrinsic::objc_sync_exit},
6362 {
"objc_arc_annotation_topdown_bbstart",
6363 llvm::Intrinsic::objc_arc_annotation_topdown_bbstart},
6364 {
"objc_arc_annotation_topdown_bbend",
6365 llvm::Intrinsic::objc_arc_annotation_topdown_bbend},
6366 {
"objc_arc_annotation_bottomup_bbstart",
6367 llvm::Intrinsic::objc_arc_annotation_bottomup_bbstart},
6368 {
"objc_arc_annotation_bottomup_bbend",
6369 llvm::Intrinsic::objc_arc_annotation_bottomup_bbend}};
6371 for (
auto &
I : RuntimeFuncs)
6372 UpgradeToIntrinsic(
I.first,
I.second);
6376 NamedMDNode *ModFlags = M.getModuleFlagsMetadata();
6380 bool HasObjCFlag =
false, HasClassProperties =
false,
Changed =
false;
6381 bool HasSwiftVersionFlag =
false;
6382 uint8_t SwiftMajorVersion, SwiftMinorVersion;
6389 if (
Op->getNumOperands() != 3)
6403 if (
ID->getString() ==
"Objective-C Image Info Version")
6405 if (
ID->getString() ==
"Objective-C Class Properties")
6406 HasClassProperties =
true;
6408 if (
ID->getString() ==
"PIC Level") {
6409 if (
auto *Behavior =
6411 uint64_t V = Behavior->getLimitedValue();
6417 if (
ID->getString() ==
"PIE Level")
6418 if (
auto *Behavior =
6425 if (
ID->getString() ==
"branch-target-enforcement" ||
6426 ID->getString().starts_with(
"sign-return-address")) {
6427 if (
auto *Behavior =
6433 Op->getOperand(1),
Op->getOperand(2)};
6443 if (
ID->getString() ==
"Objective-C Image Info Section") {
6446 Value->getString().split(ValueComp,
" ");
6447 if (ValueComp.
size() != 1) {
6448 std::string NewValue;
6449 for (
auto &S : ValueComp)
6450 NewValue += S.str();
6461 if (
ID->getString() ==
"Objective-C Garbage Collection") {
6464 assert(Md->getValue() &&
"Expected non-empty metadata");
6465 auto Type = Md->getValue()->getType();
6468 unsigned Val = Md->getValue()->getUniqueInteger().getZExtValue();
6469 if ((Val & 0xff) != Val) {
6470 HasSwiftVersionFlag =
true;
6471 SwiftABIVersion = (Val & 0xff00) >> 8;
6472 SwiftMajorVersion = (Val & 0xff000000) >> 24;
6473 SwiftMinorVersion = (Val & 0xff0000) >> 16;
6484 if (
ID->getString() ==
"amdgpu_code_object_version") {
6487 MDString::get(M.getContext(),
"amdhsa_code_object_version"),
6499 if (HasObjCFlag && !HasClassProperties) {
6505 if (HasSwiftVersionFlag) {
6509 ConstantInt::get(Int8Ty, SwiftMajorVersion));
6511 ConstantInt::get(Int8Ty, SwiftMinorVersion));
6519 NamedMDNode *CFIConsts = M.getNamedMetadata(
"cfi.functions");
6523 auto MatchesVersion = [](
const MDNode *
Op) {
6524 return Op->getNumOperands() >= 3 &&
6538 assert(!MatchesVersion(
Op) &&
"Unexpected mix of CFIConstant formats");
6539 assert(
Op->getNumOperands() >= 2 &&
6540 "Expected at least 2 operands - name and linkage type");
6552 for (
unsigned J = 2, EJ =
Op->getNumOperands(); J != EJ; ++J)
6563 auto TrimSpaces = [](
StringRef Section) -> std::string {
6565 Section.split(Components,
',');
6570 for (
auto Component : Components)
6571 OS <<
',' << Component.trim();
6576 for (
auto &GV : M.globals()) {
6577 if (!GV.hasSection())
6582 if (!Section.starts_with(
"__DATA, __objc_catlist"))
6587 GV.setSection(TrimSpaces(Section));
6603struct StrictFPUpgradeVisitor :
public InstVisitor<StrictFPUpgradeVisitor> {
6604 StrictFPUpgradeVisitor() =
default;
6607 if (!
Call.isStrictFP())
6613 Call.removeFnAttr(Attribute::StrictFP);
6614 Call.addFnAttr(Attribute::NoBuiltin);
6619struct AMDGPUUnsafeFPAtomicsUpgradeVisitor
6620 :
public InstVisitor<AMDGPUUnsafeFPAtomicsUpgradeVisitor> {
6621 AMDGPUUnsafeFPAtomicsUpgradeVisitor() =
default;
6623 void visitAtomicRMWInst(AtomicRMWInst &RMW) {
6638 if (!
F.isDeclaration() && !
F.hasFnAttribute(Attribute::StrictFP)) {
6639 StrictFPUpgradeVisitor SFPV;
6644 F.removeRetAttrs(AttributeFuncs::typeIncompatible(
6645 F.getReturnType(),
F.getAttributes().getRetAttrs()));
6646 for (
auto &Arg :
F.args())
6648 AttributeFuncs::typeIncompatible(Arg.getType(), Arg.getAttributes()));
6650 bool AddingAttrs =
false, RemovingAttrs =
false;
6651 AttrBuilder AttrsToAdd(
F.getContext());
6656 if (
Attribute A =
F.getFnAttribute(
"implicit-section-name");
6657 A.isValid() &&
A.isStringAttribute()) {
6658 F.setSection(
A.getValueAsString());
6660 RemovingAttrs =
true;
6664 A.isValid() &&
A.isStringAttribute()) {
6667 AddingAttrs = RemovingAttrs =
true;
6670 if (
Attribute A =
F.getFnAttribute(
"uniform-work-group-size");
6671 A.isValid() &&
A.isStringAttribute() && !
A.getValueAsString().empty()) {
6673 RemovingAttrs =
true;
6674 if (
A.getValueAsString() ==
"true") {
6675 AttrsToAdd.addAttribute(
"uniform-work-group-size");
6684 if (
Attribute A =
F.getFnAttribute(
"amdgpu-unsafe-fp-atomics");
6687 if (
A.getValueAsBool()) {
6688 AMDGPUUnsafeFPAtomicsUpgradeVisitor Visitor;
6694 AttrsToRemove.
addAttribute(
"amdgpu-unsafe-fp-atomics");
6695 RemovingAttrs =
true;
6702 bool HandleDenormalMode =
false;
6704 if (
Attribute Attr =
F.getFnAttribute(
"denormal-fp-math"); Attr.isValid()) {
6707 DenormalFPMath = ParsedMode;
6709 AddingAttrs = RemovingAttrs =
true;
6710 HandleDenormalMode =
true;
6714 if (
Attribute Attr =
F.getFnAttribute(
"denormal-fp-math-f32");
6718 DenormalFPMathF32 = ParsedMode;
6720 AddingAttrs = RemovingAttrs =
true;
6721 HandleDenormalMode =
true;
6725 if (HandleDenormalMode)
6726 AttrsToAdd.addDenormalFPEnvAttr(
6730 F.removeFnAttrs(AttrsToRemove);
6733 F.addFnAttrs(AttrsToAdd);
6739 if (!
F.hasFnAttribute(FnAttrName))
6740 F.addFnAttr(FnAttrName,
Value);
6747 if (!
F.hasFnAttribute(FnAttrName)) {
6749 F.addFnAttr(FnAttrName);
6751 auto A =
F.getFnAttribute(FnAttrName);
6752 if (
"false" ==
A.getValueAsString())
6753 F.removeFnAttr(FnAttrName);
6754 else if (
"true" ==
A.getValueAsString()) {
6755 F.removeFnAttr(FnAttrName);
6756 F.addFnAttr(FnAttrName);
6762 Triple T(M.getTargetTriple());
6763 if (!
T.isThumb() && !
T.isARM() && !
T.isAArch64())
6773 NamedMDNode *ModFlags = M.getModuleFlagsMetadata();
6777 if (
Op->getNumOperands() != 3)
6786 uint64_t *ValPtr = IDStr ==
"branch-target-enforcement" ? &BTEValue
6787 : IDStr ==
"branch-protection-pauth-lr" ? &BPPLRValue
6788 : IDStr ==
"guarded-control-stack" ? &GCSValue
6789 : IDStr ==
"sign-return-address" ? &SRAValue
6790 : IDStr ==
"sign-return-address-all" ? &SRAALLValue
6791 : IDStr ==
"sign-return-address-with-bkey"
6797 *ValPtr = CI->getZExtValue();
6803 bool BTE = BTEValue == 1;
6804 bool BPPLR = BPPLRValue == 1;
6805 bool GCS = GCSValue == 1;
6806 bool SRA = SRAValue == 1;
6809 if (SRA && SRAALLValue == 1)
6810 SignTypeValue =
"all";
6813 if (SRA && SRABKeyValue == 1)
6814 SignKeyValue =
"b_key";
6816 for (
Function &
F : M.getFunctionList()) {
6817 if (
F.isDeclaration())
6824 if (
auto A =
F.getFnAttribute(
"sign-return-address");
6825 A.isValid() &&
"none" ==
A.getValueAsString()) {
6826 F.removeFnAttr(
"sign-return-address");
6827 F.removeFnAttr(
"sign-return-address-key");
6843 if (SRAALLValue == 1)
6845 if (SRABKeyValue == 1)
6854 if (
T->getNumOperands() < 1)
6859 return S->getString().starts_with(
"llvm.vectorizer.");
6863 StringRef OldPrefix =
"llvm.vectorizer.";
6866 if (OldTag ==
"llvm.vectorizer.unroll")
6878 if (
T->getNumOperands() < 1)
6883 if (!OldTag->getString().starts_with(
"llvm.vectorizer."))
6888 Ops.reserve(
T->getNumOperands());
6890 for (
unsigned I = 1,
E =
T->getNumOperands();
I !=
E; ++
I)
6891 Ops.push_back(
T->getOperand(
I));
6905 Ops.reserve(
T->getNumOperands());
6916 if ((
T.isSPIR() || (
T.isSPIRV() && !
T.isSPIRVLogical())) &&
6917 !
DL.contains(
"-G") && !
DL.starts_with(
"G")) {
6918 return DL.empty() ? std::string(
"G1") : (
DL +
"-G1").str();
6921 if (
T.isLoongArch64() ||
T.isRISCV64()) {
6923 auto I =
DL.find(
"-n64-");
6925 return (
DL.take_front(
I) +
"-n32:64-" +
DL.drop_front(
I + 5)).str();
6930 std::string Res =
DL.str();
6933 if (!
DL.contains(
"-G") && !
DL.starts_with(
"G"))
6934 Res.append(Res.empty() ?
"G1" :
"-G1");
6942 if (!
DL.contains(
"-ni") && !
DL.starts_with(
"ni"))
6943 Res.append(
"-ni:7:8:9");
6945 if (
DL.ends_with(
"ni:7"))
6947 if (
DL.ends_with(
"ni:7:8"))
6952 if (!
DL.contains(
"-p7") && !
DL.starts_with(
"p7"))
6953 Res.append(
"-p7:160:256:256:32");
6954 if (!
DL.contains(
"-p8") && !
DL.starts_with(
"p8"))
6955 Res.append(
"-p8:128:128:128:48");
6956 constexpr StringRef OldP8(
"-p8:128:128-");
6957 if (
DL.contains(OldP8))
6958 Res.replace(Res.find(OldP8), OldP8.
size(),
"-p8:128:128:128:48-");
6959 if (!
DL.contains(
"-p9") && !
DL.starts_with(
"p9"))
6960 Res.append(
"-p9:192:256:256:32");
6964 if (!
DL.contains(
"m:e"))
6965 Res = Res.empty() ?
"m:e" :
"m:e-" + Res;
6970 if (
T.isSystemZ() && !
DL.empty()) {
6972 if (!
DL.contains(
"-S64"))
6973 return "E-S64" +
DL.drop_front(1).str();
6977 auto AddPtr32Ptr64AddrSpaces = [&
DL, &Res]() {
6980 StringRef AddrSpaces{
"-p270:32:32-p271:32:32-p272:64:64"};
6981 if (!
DL.contains(AddrSpaces)) {
6983 Regex R(
"^([Ee]-m:[a-z](-p:32:32)?)(-.*)$");
6984 if (R.match(Res, &
Groups))
6990 if (
T.isAArch64()) {
6992 if (!
DL.empty() && !
DL.contains(
"-Fn32"))
6993 Res.append(
"-Fn32");
6994 AddPtr32Ptr64AddrSpaces();
6998 if (
T.isSPARC() || (
T.isMIPS64() && !
DL.contains(
"m:m")) ||
T.isPPC64() ||
7002 std::string I64 =
"-i64:64";
7003 std::string I128 =
"-i128:128";
7005 size_t Pos = Res.find(I64);
7006 if (Pos !=
size_t(-1))
7007 Res.insert(Pos + I64.size(), I128);
7011 if (
T.isPPC() &&
T.isOSAIX() && !
DL.contains(
"f64:32:64") && !
DL.empty()) {
7012 size_t Pos = Res.find(
"-S128");
7015 Res.insert(Pos,
"-f64:32:64");
7021 AddPtr32Ptr64AddrSpaces();
7029 if (!
T.isOSIAMCU()) {
7030 std::string I128 =
"-i128:128";
7033 Regex R(
"^(e(-[mpi][^-]*)*)((-[^mpi][^-]*)*)$");
7034 if (R.match(Res, &
Groups))
7042 if (
T.isWindowsMSVCEnvironment() && !
T.isArch64Bit()) {
7044 auto I =
Ref.find(
"-f80:32-");
7046 Res = (
Ref.take_front(
I) +
"-f80:128-" +
Ref.drop_front(
I + 8)).str();
7054 Attribute A =
B.getAttribute(
"no-frame-pointer-elim");
7057 FramePointer =
A.getValueAsString() ==
"true" ?
"all" :
"none";
7058 B.removeAttribute(
"no-frame-pointer-elim");
7060 if (
B.contains(
"no-frame-pointer-elim-non-leaf")) {
7062 if (FramePointer !=
"all")
7063 FramePointer =
"non-leaf";
7064 B.removeAttribute(
"no-frame-pointer-elim-non-leaf");
7066 if (!FramePointer.
empty())
7067 B.addAttribute(
"frame-pointer", FramePointer);
7069 A =
B.getAttribute(
"null-pointer-is-valid");
7072 bool NullPointerIsValid =
A.getValueAsString() ==
"true";
7073 B.removeAttribute(
"null-pointer-is-valid");
7074 if (NullPointerIsValid)
7075 B.addAttribute(Attribute::NullPointerIsValid);
7078 A =
B.getAttribute(
"uniform-work-group-size");
7082 bool IsTrue = Val ==
"true";
7083 B.removeAttribute(
"uniform-work-group-size");
7085 B.addAttribute(
"uniform-work-group-size");
7096 return OBD.
getTag() ==
"clang.arc.attachedcall" &&
assert(UImm &&(UImm !=~static_cast< T >(0)) &&"Invalid immediate!")
AMDGPU address space definition.
AMDGPU Register Bank Select
MachineBasicBlock MachineBasicBlock::iterator DebugLoc DL
This file contains the simple types necessary to represent the attributes associated with functions a...
static bool upgradeIntrinsicDeclWithDefaultArgs(Function *F, Function *&NewFn)
static Value * upgradeX86VPERMT2Intrinsics(IRBuilder<> &Builder, CallBase &CI, bool ZeroMask, bool IndexForm)
static Metadata * upgradeLoopArgument(Metadata *MD)
static bool isXYZ(StringRef S)
static bool upgradeIntrinsicFunction1(Function *F, Function *&NewFn, bool CanUpgradeDebugIntrinsicsToRecords)
static Value * upgradeX86PSLLDQIntrinsics(IRBuilder<> &Builder, Value *Op, unsigned Shift)
static Intrinsic::ID shouldUpgradeNVPTXSharedClusterIntrinsic(Function *F, StringRef Name)
static bool upgradeRetainReleaseMarker(Module &M)
This checks for objc retain release marker which should be upgraded.
static Value * upgradeX86vpcom(IRBuilder<> &Builder, CallBase &CI, unsigned Imm, bool IsSigned)
static Value * upgradeMaskToInt(IRBuilder<> &Builder, CallBase &CI)
static bool convertIntrinsicValidType(StringRef Name, const FunctionType *FuncTy)
static Value * upgradeX86Rotate(IRBuilder<> &Builder, CallBase &CI, bool IsRotateRight)
static bool upgradeX86MultiplyAddBytes(Function *F, Intrinsic::ID IID, Function *&NewFn)
static void setFunctionAttrIfNotSet(Function &F, StringRef FnAttrName, StringRef Value)
static Intrinsic::ID shouldUpgradeNVPTXBF16Intrinsic(StringRef Name)
static bool upgradeSingleNVVMAnnotation(GlobalValue *GV, StringRef K, const Metadata *V)
static MDNode * unwrapMAVOp(CallBase *CI, unsigned Op)
Helper to unwrap intrinsic call MetadataAsValue operands.
static MDString * upgradeLoopTag(LLVMContext &C, StringRef OldTag)
static void upgradeNVVMFnVectorAttr(const StringRef Attr, const char DimC, GlobalValue *GV, const Metadata *V)
static bool upgradeX86MaskedFPCompare(Function *F, Intrinsic::ID IID, Function *&NewFn)
static Value * upgradeX86ALIGNIntrinsics(IRBuilder<> &Builder, Value *Op0, Value *Op1, Value *Shift, Value *Passthru, Value *Mask, bool IsVALIGN)
static Value * upgradeAbs(IRBuilder<> &Builder, CallBase &CI)
static Value * emitX86Select(IRBuilder<> &Builder, Value *Mask, Value *Op0, Value *Op1)
static Value * upgradeAArch64IntrinsicCall(StringRef Name, CallBase *CI, Function *F, IRBuilder<> &Builder)
static Value * upgradeMaskedMove(IRBuilder<> &Builder, CallBase &CI)
static bool upgradeX86IntrinsicFunction(Function *F, StringRef Name, Function *&NewFn)
static Value * applyX86MaskOn1BitsVec(IRBuilder<> &Builder, Value *Vec, Value *Mask)
static bool consumeNVVMPtrAddrSpace(StringRef &Name)
static bool shouldUpgradeX86Intrinsic(Function *F, StringRef Name)
static Value * upgradeX86PSRLDQIntrinsics(IRBuilder<> &Builder, Value *Op, unsigned Shift)
static Intrinsic::ID shouldUpgradeNVPTXTMAG2SIntrinsics(Function *F, StringRef Name)
static bool isOldLoopArgument(Metadata *MD)
static Value * upgradeARMIntrinsicCall(StringRef Name, CallBase *CI, Function *F, IRBuilder<> &Builder)
static bool upgradeX86IntrinsicsWith8BitMask(Function *F, Intrinsic::ID IID, Function *&NewFn)
static Value * upgradeVectorSplice(CallBase *CI, IRBuilder<> &Builder)
static Value * upgradeAMDGCNIntrinsicCall(StringRef Name, CallBase *CI, Function *F, IRBuilder<> &Builder)
static Value * upgradeMaskedLoad(IRBuilder<> &Builder, Value *Ptr, Value *Passthru, Value *Mask, bool Aligned)
static Metadata * unwrapMAVMetadataOp(CallBase *CI, unsigned Op)
Helper to unwrap Metadata MetadataAsValue operands, such as the Value field.
static bool upgradeX86BF16Intrinsic(Function *F, Intrinsic::ID IID, Function *&NewFn)
static bool upgradeArmOrAarch64IntrinsicFunction(bool IsArm, Function *F, StringRef Name, Function *&NewFn)
static bool upgradeIntrinsicCallWithDefaultArgs(CallBase *CI, Function *NewFn, IRBuilder<> &Builder)
static Value * getX86MaskVec(IRBuilder<> &Builder, Value *Mask, unsigned NumElts)
static Value * emitX86ScalarSelect(IRBuilder<> &Builder, Value *Mask, Value *Op0, Value *Op1)
static Value * upgradeX86ConcatShift(IRBuilder<> &Builder, CallBase &CI, bool IsShiftRight, bool ZeroMask)
static void rename(GlobalValue *GV)
static bool upgradePTESTIntrinsic(Function *F, Intrinsic::ID IID, Function *&NewFn)
static bool upgradeX86BF16DPIntrinsic(Function *F, Intrinsic::ID IID, Function *&NewFn)
static cl::opt< bool > DisableAutoUpgradeDebugInfo("disable-auto-upgrade-debug-info", cl::desc("Disable autoupgrade of debug info"))
static Value * upgradeMaskedCompare(IRBuilder<> &Builder, CallBase &CI, unsigned CC, bool Signed)
static Value * upgradeX86BinaryIntrinsics(IRBuilder<> &Builder, CallBase &CI, Intrinsic::ID IID)
static Value * upgradeNVVMIntrinsicCall(StringRef Name, CallBase *CI, Function *F, IRBuilder<> &Builder)
static Value * upgradeX86MaskedShift(IRBuilder<> &Builder, CallBase &CI, Intrinsic::ID IID)
static bool upgradeAVX512MaskToSelect(StringRef Name, IRBuilder<> &Builder, CallBase &CI, Value *&Rep)
static void upgradeDbgIntrinsicToDbgRecord(StringRef Name, CallBase *CI)
Convert debug intrinsic calls to non-instruction debug records.
static void ConvertFunctionAttr(Function &F, bool Set, StringRef FnAttrName)
static Value * upgradePMULDQ(IRBuilder<> &Builder, CallBase &CI, bool IsSigned)
static void reportFatalUsageErrorWithCI(StringRef reason, CallBase *CI)
static Value * upgradeMaskedStore(IRBuilder<> &Builder, Value *Ptr, Value *Data, Value *Mask, bool Aligned)
static Value * upgradeConvertIntrinsicCall(StringRef Name, CallBase *CI, Function *F, IRBuilder<> &Builder)
static bool upgradeX86MultiplyAddWords(Function *F, Intrinsic::ID IID, Function *&NewFn)
static Value * upgradeX86IntrinsicCall(StringRef Name, CallBase *CI, Function *F, IRBuilder<> &Builder)
static GCRegistry::Add< ErlangGC > A("erlang", "erlang-compatible garbage collector")
static GCRegistry::Add< CoreCLRGC > E("coreclr", "CoreCLR-compatible GC")
static GCRegistry::Add< OcamlGC > B("ocaml", "ocaml 3.10-compatible GC")
This file contains the declarations for the subclasses of Constant, which represent the different fla...
This file contains constants used for implementing Dwarf debug support.
Module.h This file contains the declarations for the Module class.
const AbstractManglingParser< Derived, Alloc >::OperatorInfo AbstractManglingParser< Derived, Alloc >::Ops[]
static bool isZero(Value *V, const DataLayout &DL, DominatorTree *DT, AssumptionCache *AC)
NVPTX address space definition.
static unsigned getNumElements(Type *Ty)
static bool contains(SmallPtrSetImpl< ConstantExpr * > &Cache, ConstantExpr *Expr, Constant *C)
This file implements the StringSwitch template, which mimics a switch() statement whose cases are str...
static SymbolRef::Type getType(const Symbol *Sym)
LocallyHashedType DenseMapInfo< LocallyHashedType >::Empty
static const X86InstrFMA3Group Groups[]
Class for arbitrary precision integers.
Represent a constant reference to an array (0 or more elements consecutively in memory),...
Class to represent array types.
static LLVM_ABI ArrayType * get(Type *ElementType, uint64_t NumElements)
This static method is the primary way to construct an ArrayType.
Type * getElementType() const
an instruction that atomically reads a memory location, combines it with another value,...
void setVolatile(bool V)
Specify whether this is a volatile RMW or not.
BinOp
This enumeration lists the possible modifications atomicrmw can make.
@ USubCond
Subtract only if no unsigned overflow.
@ Min
*p = old <signed v ? old : v
@ USubSat
*p = usub.sat(old, v) usub.sat matches the behavior of llvm.usub.sat.
@ UIncWrap
Increment one up to a maximum value.
@ Max
*p = old >signed v ? old : v
@ FMin
*p = minnum(old, v) minnum matches the behavior of llvm.minnum.
@ FMax
*p = maxnum(old, v) maxnum matches the behavior of llvm.maxnum.
@ UDecWrap
Decrement one until a minimum value or zero.
bool isFloatingPointOperation() const
This class stores enough information to efficiently remove some attributes from an existing AttrBuild...
AttributeMask & addAttribute(Attribute::AttrKind Val)
Add an attribute to the mask.
Functions, function parameters, and return types can have attributes to indicate how they should be t...
static LLVM_ABI Attribute getWithStackAlignment(LLVMContext &Context, Align Alignment)
static LLVM_ABI Attribute get(LLVMContext &Context, AttrKind Kind, uint64_t Val=0)
Return a uniquified Attribute object.
Base class for all callable instructions (InvokeInst and CallInst) Holds everything related to callin...
void setCallingConv(CallingConv::ID CC)
LLVM_ABI void getOperandBundlesAsDefs(SmallVectorImpl< OperandBundleDef > &Defs) const
Return the list of operand bundles attached to this instruction as a vector of OperandBundleDefs.
Function * getCalledFunction() const
Returns the function called, or null if this is an indirect function invocation or the function signa...
CallingConv::ID getCallingConv() const
Value * getCalledOperand() const
void setAttributes(AttributeList A)
Set the attributes for this call.
Value * getArgOperand(unsigned i) const
FunctionType * getFunctionType() const
LLVM_ABI Intrinsic::ID getIntrinsicID() const
Returns the intrinsic ID of the intrinsic called or Intrinsic::not_intrinsic if the called function i...
iterator_range< User::op_iterator > args()
Iteration adapter for range-for loops.
void setCalledOperand(Value *V)
unsigned arg_size() const
AttributeList getAttributes() const
Return the attributes for this call.
void setCalledFunction(Function *Fn)
Sets the function called, including updating the function type.
This class represents a function call, abstracting a target machine's calling convention.
void setTailCallKind(TailCallKind TCK)
static LLVM_ABI CastInst * Create(Instruction::CastOps, Value *S, Type *Ty, const Twine &Name="", InsertPosition InsertBefore=nullptr)
Provides a way to construct any of the CastInst subclasses using an opcode instead of the subclass's ...
static LLVM_ABI bool castIsValid(Instruction::CastOps op, Type *SrcTy, Type *DstTy)
This method can be used to determine if a cast from SrcTy to DstTy using Opcode op is valid or not.
Predicate
This enumeration lists the possible predicates for CmpInst subclasses.
@ ICMP_SLT
signed less than
@ ICMP_SLE
signed less or equal
@ ICMP_UGE
unsigned greater or equal
@ ICMP_UGT
unsigned greater than
@ ICMP_SGT
signed greater than
@ ICMP_ULT
unsigned less than
@ ICMP_SGE
signed greater or equal
@ ICMP_ULE
unsigned less or equal
static LLVM_ABI ConstantAggregateZero * get(Type *Ty)
static LLVM_ABI Constant * get(ArrayType *T, ArrayRef< Constant * > V)
static LLVM_ABI Constant * getIntToPtr(Constant *C, Type *Ty, bool OnlyIfReduced=false)
static LLVM_ABI Constant * getPointerCast(Constant *C, Type *Ty)
Create a BitCast, AddrSpaceCast, or a PtrToInt cast constant expression.
static LLVM_ABI Constant * getPtrToInt(Constant *C, Type *Ty, bool OnlyIfReduced=false)
This is the shared class of boolean and integer constants.
bool isZero() const
This is just a convenience method to make client code smaller for a common code.
uint64_t getZExtValue() const
Return the constant as a 64-bit unsigned integer value after it has been zero extended as appropriate...
static LLVM_ABI ConstantPointerNull * get(PointerType *T)
Static factory methods - Return objects of the specified value.
static LLVM_ABI Constant * get(StructType *T, ArrayRef< Constant * > V)
static LLVM_ABI ConstantTokenNone * get(LLVMContext &Context)
Return the ConstantTokenNone.
This is an important base class in LLVM.
static LLVM_ABI Constant * getAllOnesValue(Type *Ty)
static LLVM_ABI Constant * getNullValue(Type *Ty)
Constructor to create a '0' constant of arbitrary type.
static LLVM_ABI DIExpression * append(const DIExpression *Expr, ArrayRef< uint64_t > Ops)
Append the opcodes Ops to DIExpr.
A parsed version of the target data layout string in and methods for querying it.
static LLVM_ABI DbgLabelRecord * createUnresolvedDbgLabelRecord(MDNode *Label)
For use during parsing; creates a DbgLabelRecord from as-of-yet unresolved MDNodes.
Base class for non-instruction debug metadata records that have positions within IR.
void setDebugLoc(DebugLoc Loc)
static LLVM_ABI DbgVariableRecord * createUnresolvedDbgVariableRecord(LocationType Type, Metadata *Val, MDNode *Variable, MDNode *Expression, MDNode *AssignID, Metadata *Address, MDNode *AddressExpression)
Used to create DbgVariableRecords during parsing, where some metadata references may still be unresol...
Convenience struct for specifying and reasoning about fast-math flags.
void setApproxFunc(bool B=true)
static LLVM_ABI FixedVectorType * get(Type *ElementType, unsigned NumElts)
Class to represent function types.
Type * getParamType(unsigned i) const
Parameter type accessors.
Type * getReturnType() const
static LLVM_ABI FunctionType * get(Type *Result, ArrayRef< Type * > Params, bool isVarArg)
This static method is the primary way of constructing a FunctionType.
static Function * Create(FunctionType *Ty, LinkageTypes Linkage, unsigned AddrSpace, const Twine &N="", Module *M=nullptr)
FunctionType * getFunctionType() const
Returns the FunctionType for me.
Intrinsic::ID getIntrinsicID() const LLVM_READONLY
getIntrinsicID - This method returns the ID number of the specified function, or Intrinsic::not_intri...
const Function & getFunction() const
void eraseFromParent()
eraseFromParent - This method unlinks 'this' from the containing module and deletes it.
Type * getReturnType() const
Returns the type of the ret val.
Argument * getArg(unsigned i) const
static LLVM_ABI GUID getGUIDAssumingExternalLinkage(StringRef GlobalName)
Return a 64-bit global unique ID constructed from the name of a global symbol.
LinkageTypes getLinkage() const
uint64_t GUID
Declare a type to represent a global unique identifier for a global value.
static StringRef dropLLVMManglingEscape(StringRef Name)
If the given string begins with the GlobalValue name mangling escape character '\1',...
Type * getValueType() const
const Constant * getInitializer() const
getInitializer - Return the initializer for this global variable.
bool hasInitializer() const
Definitions have initializers, declarations don't.
PointerType * getPtrTy(unsigned AddrSpace=0)
Fetch the type representing a pointer.
This provides a uniform API for creating instructions and inserting them into a basic block: either a...
Base class for instruction visitors.
const DebugLoc & getDebugLoc() const
Return the debug location for this node as a DebugLoc.
LLVM_ABI const Module * getModule() const
Return the module owning the function this instruction belongs to or nullptr it the function does not...
LLVM_ABI InstListType::iterator eraseFromParent()
This method unlinks 'this' from the containing basic block and deletes it.
LLVM_ABI void setMetadata(unsigned KindID, MDNode *Node)
Set the metadata of the specified kind to the specified node.
LLVM_ABI FastMathFlags getFastMathFlags() const LLVM_READONLY
Convenience function for getting all the fast-math flags, which must be an operator which supports th...
LLVM_ABI void copyMetadata(const Instruction &SrcInst, ArrayRef< unsigned > WL=ArrayRef< unsigned >())
Copy metadata from SrcInst to this instruction.
LLVM_ABI const DataLayout & getDataLayout() const
Get the data layout of the module this instruction belongs to.
This is an important class for using LLVM in a threaded context.
LLVM_ABI SyncScope::ID getOrInsertSyncScopeID(StringRef SSN)
getOrInsertSyncScopeID - Maps synchronization scope name to synchronization scope ID.
An instruction for reading from memory.
LLVM_ABI MDNode * createRange(const APInt &Lo, const APInt &Hi)
Return metadata describing the range [Lo, Hi).
const MDOperand & getOperand(unsigned I) const
static MDTuple * get(LLVMContext &Context, ArrayRef< Metadata * > MDs)
unsigned getNumOperands() const
Return number of MDNode operands.
LLVMContext & getContext() const
Tracking metadata reference owned by Metadata.
LLVM_ABI StringRef getString() const
static LLVM_ABI MDString * get(LLVMContext &Context, StringRef Str)
static MDTuple * get(LLVMContext &Context, ArrayRef< Metadata * > MDs)
A Module instance is used to store all the information related to an LLVM module.
ModFlagBehavior
This enumeration defines the supported behaviors of module flags.
@ Override
Uses the specified value, regardless of the behavior or value of the other module.
@ Error
Emits an error if two values disagree, otherwise the resulting value is that of the operands.
@ Min
Takes the min of the two values, which are required to be integers.
@ Max
Takes the max of the two values, which are required to be integers.
LLVM_ABI void setOperand(unsigned I, MDNode *New)
LLVM_ABI MDNode * getOperand(unsigned i) const
LLVM_ABI unsigned getNumOperands() const
LLVM_ABI void clearOperands()
Drop all references to this node's operands.
iterator_range< op_iterator > operands()
LLVM_ABI void addOperand(MDNode *M)
ArrayRef< InputTy > inputs() const
static LLVM_ABI PoisonValue * get(Type *T)
Static factory methods - Return an 'poison' object of the specified type.
LLVM_ABI bool match(StringRef String, SmallVectorImpl< StringRef > *Matches=nullptr, std::string *Error=nullptr) const
matches - Match the regex against a given String.
static LLVM_ABI ScalableVectorType * get(Type *ElementType, unsigned MinNumElts)
ArrayRef< int > getShuffleMask() const
std::pair< iterator, bool > insert(PtrType Ptr)
Inserts Ptr if and only if there is no element in the container equal to Ptr.
SmallPtrSet - This class implements a set which is optimized for holding SmallSize or less elements.
SmallString - A SmallString is just a SmallVector with methods and accessors that make it work better...
void append(ItTy in_start, ItTy in_end)
Add the specified range to the end of the SmallVector.
void push_back(const T &Elt)
This is a 'vector' (really, a variable-sized array), optimized for the case when the array is small.
An instruction for storing to memory.
A wrapper around a string literal that serves as a proxy for constructing global tables of StringRefs...
Represent a constant reference to a string, i.e.
std::pair< StringRef, StringRef > split(char Separator) const
Split into two substrings around the first occurrence of a separator character.
static constexpr size_t npos
constexpr StringRef substr(size_t Start, size_t N=npos) const
Return a reference to the substring from [Start, Start + N).
bool starts_with(StringRef Prefix) const
Check if this string starts with the given Prefix.
constexpr bool empty() const
Check if the string is empty.
StringRef drop_front(size_t N=1) const
Return a StringRef equal to 'this' but with the first N elements dropped.
constexpr size_t size() const
Get the string size.
StringRef trim(char Char) const
Return string with consecutive Char characters starting from the left and right removed.
A switch()-like statement whose cases are string literals.
StringSwitch & Case(StringLiteral S, T Value)
StringSwitch & StartsWith(StringLiteral S, T Value)
StringSwitch & Cases(std::initializer_list< StringLiteral > CaseStrings, T Value)
Class to represent struct types.
static LLVM_ABI StructType * get(LLVMContext &Context, ArrayRef< Type * > Elements, bool isPacked=false)
This static method is the primary way to create a literal StructType.
unsigned getNumElements() const
Random access to the elements.
Type * getElementType(unsigned N) const
The TimeTraceScope is a helper class to call the begin and end functions of the time trace profiler.
Triple - Helper class for working with autoconf configuration names.
Twine - A lightweight data structure for efficiently representing the concatenation of temporary valu...
The instances of the Type class are immutable: once they are created, they are never changed.
static LLVM_ABI IntegerType * getInt64Ty(LLVMContext &C)
bool isVectorTy() const
True if this is an instance of VectorType.
static LLVM_ABI IntegerType * getInt32Ty(LLVMContext &C)
bool isFloatTy() const
Return true if this is 'float', a 32-bit IEEE fp type.
bool isBFloatTy() const
Return true if this is 'bfloat', a 16-bit bfloat type.
LLVM_ABI unsigned getPointerAddressSpace() const
Get the address space of this pointer or pointer vector type.
static LLVM_ABI IntegerType * getInt8Ty(LLVMContext &C)
Type * getScalarType() const
If this is a vector type, return the element type, otherwise return 'this'.
LLVM_ABI TypeSize getPrimitiveSizeInBits() const LLVM_READONLY
Return the basic size of this type if it is a primitive type.
static LLVM_ABI IntegerType * getInt16Ty(LLVMContext &C)
LLVM_ABI unsigned getScalarSizeInBits() const LLVM_READONLY
If this is a vector type, return the getPrimitiveSizeInBits value for the element type.
bool isPtrOrPtrVectorTy() const
Return true if this is a pointer type or a vector of pointer types.
bool isIntegerTy() const
True if this is an instance of IntegerType.
bool isFPOrFPVectorTy() const
Return true if this is a FP type or a vector of FP.
static LLVM_ABI Type * getFloatTy(LLVMContext &C)
static LLVM_ABI Type * getBFloatTy(LLVMContext &C)
static LLVM_ABI Type * getHalfTy(LLVMContext &C)
Value * getOperand(unsigned i) const
unsigned getNumOperands() const
LLVM Value Representation.
Type * getType() const
All values are typed, get the type of this value.
LLVM_ABI void print(raw_ostream &O, bool IsForDebug=false) const
Implement operator<< on Value.
LLVM_ABI void setName(const Twine &Name)
Change the name of the value.
LLVM_ABI void replaceAllUsesWith(Value *V)
Change all uses of this to point to a new Value.
LLVMContext & getContext() const
All values hold a context through their type.
iterator_range< user_iterator > users()
LLVM_ABI const Value * stripPointerCasts() const
Strip off pointer casts, all-zero GEPs and address space casts.
LLVM_ABI StringRef getName() const
Return a constant reference to the value's name.
LLVM_ABI void takeName(Value *V)
Transfer the name from V to this value.
Base class of all SIMD vector types.
static VectorType * getInteger(VectorType *VTy)
This static method gets a VectorType with the same number of elements as the input type,...
static LLVM_ABI VectorType * get(Type *ElementType, ElementCount EC)
This static method is the primary way to construct an VectorType.
constexpr ScalarTy getFixedValue() const
const ParentTy * getParent() const
self_iterator getIterator()
A raw_ostream that writes to an SmallVector or SmallString.
StringRef str() const
Return a StringRef for the vector contents.
#define llvm_unreachable(msg)
Marks that the current location is not supposed to be reachable.
@ LOCAL_ADDRESS
Address space for local memory.
@ FLAT_ADDRESS
Address space for flat memory.
@ PRIVATE_ADDRESS
Address space for private memory.
unsigned ID
LLVM IR allows to use arbitrary numbers as calling convention identifiers.
@ PTX_Kernel
Call to a PTX kernel. Passes all arguments in parameter space.
@ C
The default llvm calling convention, compatible with C.
LLVM_ABI std::optional< Function * > remangleIntrinsicFunction(Function *F)
LLVM_ABI Function * getOrInsertDeclaration(Module *M, ID id, ArrayRef< Type * > OverloadTys={})
Look up the Function declaration of the intrinsic id in the Module M.
LLVM_ABI ID lookupIntrinsicID(StringRef Name)
This does the actual lookup of an intrinsic ID which matches the given function name.
LLVM_ABI AttributeList getAttributes(LLVMContext &C, ID id, FunctionType *FT)
Return the attributes for an intrinsic.
LLVM_ABI bool isOverloaded(ID id)
Returns true if the intrinsic can be overloaded.
LLVM_ABI bool isSignatureValid(Intrinsic::ID ID, FunctionType *FT, SmallVectorImpl< Type * > &OverloadTys, raw_ostream &OS=nulls())
Returns true if FT is a valid function type for intrinsic ID.
LLVM_ABI bool hasStructReturnType(ID id)
Returns true if id has a struct return type.
LLVM_ABI std::pair< unsigned, ArrayRef< uint64_t > > getAllDefaultArgValues(ID IID)
Returns the first default argument index and an ArrayRef of all default values for the trailing param...
@ ADDRESS_SPACE_SHARED_CLUSTER
constexpr StringLiteral GridConstant("nvvm.grid_constant")
constexpr StringLiteral MaxNTID("nvvm.maxntid")
constexpr StringLiteral MaxNReg("nvvm.maxnreg")
constexpr StringLiteral MinCTASm("nvvm.minctasm")
constexpr StringLiteral ReqNTID("nvvm.reqntid")
constexpr StringLiteral MaxClusterRank("nvvm.maxclusterrank")
constexpr StringLiteral ClusterDim("nvvm.cluster_dim")
std::enable_if_t< detail::IsValidPointer< X, Y >::value, X * > dyn_extract_or_null(Y &&MD)
Extract a Value from Metadata, if any, allowing null.
std::enable_if_t< detail::IsValidPointer< X, Y >::value, X * > dyn_extract(Y &&MD)
Extract a Value from Metadata, if any.
std::enable_if_t< detail::IsValidPointer< X, Y >::value, X * > extract(Y &&MD)
Extract a Value from Metadata.
This is an optimization pass for GlobalISel generic memory operations.
LLVM_ABI void UpgradeIntrinsicCall(CallBase *CB, Function *NewFn)
This is the complement to the above, replacing a specific call to an intrinsic function with a call t...
LLVM_ABI void UpgradeSectionAttributes(Module &M)
auto size(R &&Range, std::enable_if_t< std::is_base_of< std::random_access_iterator_tag, typename std::iterator_traits< decltype(Range.begin())>::iterator_category >::value, void > *=nullptr)
Get the size of a range.
LLVM_ABI void UpgradeInlineAsmString(std::string *AsmStr)
Upgrade comment in call to inline asm that represents an objc retain release marker.
bool isValidAtomicOrdering(Int I)
decltype(auto) dyn_cast(const From &Val)
dyn_cast<X> - Return the argument parameter cast to the specified type.
LLVM_ABI bool UpgradeIntrinsicFunction(Function *F, Function *&NewFn, bool CanUpgradeDebugIntrinsicsToRecords=true)
This is a more granular function that simply checks an intrinsic function for upgrading,...
LLVM_ABI MDNode * upgradeInstructionLoopAttachment(MDNode &N)
Upgrade the loop attachment metadata node.
auto dyn_cast_if_present(const Y &Val)
dyn_cast_if_present<X> - Functionally identical to dyn_cast, except that a null (or none in the case ...
LLVM_ABI void UpgradeAttributes(AttrBuilder &B)
Upgrade attributes that changed format or kind.
LLVM_ABI void UpgradeCallsToIntrinsic(Function *F)
This is an auto-upgrade hook for any old intrinsic function syntaxes which need to have both the func...
LLVM_ABI void UpgradeNVVMAnnotations(Module &M)
Convert legacy nvvm.annotations metadata to appropriate function attributes.
iterator_range< early_inc_iterator_impl< detail::IterOfRange< RangeT > > > make_early_inc_range(RangeT &&Range)
Make a range that does early increment to allow mutation of the underlying range without disrupting i...
LLVM_ABI bool UpgradeModuleFlags(Module &M)
This checks for module flags which should be upgraded.
std::string utostr(uint64_t X, bool isNeg=false)
constexpr bool isPowerOf2_64(uint64_t Value)
Return true if the argument is a power of two > 0 (64 bit edition.)
LLVM_ABI bool UpgradeCFIFunctionsMetadata(Module &M)
Upgrade the cfi.functions metadata node by calculating and inserting the GUID for each function entry...
LLVM_ABI void copyModuleAttrToFunctions(Module &M)
Copies module attributes to the functions in the module.
LLVM_ABI void UpgradeOperandBundles(std::vector< OperandBundleDef > &OperandBundles)
Upgrade operand bundles (without knowing about their user instruction).
LLVM_ABI Constant * UpgradeBitCastExpr(unsigned Opc, Constant *C, Type *DestTy)
This is an auto-upgrade for bitcast constant expression between pointers with different address space...
auto dyn_cast_or_null(const Y &Val)
constexpr bool isPowerOf2_32(uint32_t Value)
Return true if the argument is a power of two > 0.
LLVM_ABI raw_ostream & dbgs()
dbgs() - This returns a reference to a raw_ostream for debugging messages.
LLVM_ABI std::string UpgradeDataLayoutString(StringRef DL, StringRef Triple)
Upgrade the datalayout string by adding a section for address space pointers.
bool none_of(R &&Range, UnaryPredicate P)
Provide wrappers to std::none_of which take ranges instead of having to pass begin/end explicitly.
LLVM_ABI void report_fatal_error(Error Err, bool gen_crash_diag=true)
bool isa(const From &Val)
isa<X> - Return true if the parameter to the template is an instance of one of the template type argu...
LLVM_ABI GlobalVariable * UpgradeGlobalVariable(GlobalVariable *GV)
This checks for global variables which should be upgraded.
LLVM_ABI raw_fd_ostream & errs()
This returns a reference to a raw_ostream for standard error.
LLVM_ABI bool StripDebugInfo(Module &M)
Strip debug info in the module if it exists.
AtomicOrdering
Atomic ordering for LLVM's memory model.
@ Ref
The access may reference the value stored in memory.
std::string join(IteratorT Begin, IteratorT End, StringRef Separator)
Joins the strings in the range [Begin, End), adding Separator between the elements.
OperandBundleDefT< Value * > OperandBundleDef
LLVM_ABI Instruction * UpgradeBitCastInst(unsigned Opc, Value *V, Type *DestTy, Instruction *&Temp)
This is an auto-upgrade for bitcast between pointers with different address spaces: the instruction i...
DWARFExpression::Operation Op
@ Dynamic
Denotes mode unknown at compile time.
ArrayRef(const T &OneElt) -> ArrayRef< T >
DenormalMode parseDenormalFPAttribute(StringRef Str)
Returns the denormal mode to use for inputs and outputs.
decltype(auto) cast(const From &Val)
cast<X> - Return the argument parameter cast to the specified type.
auto find_if(R &&Range, UnaryPredicate P)
Provide wrappers to std::find_if which take ranges instead of having to pass begin/end explicitly.
void erase_if(Container &C, UnaryPredicate P)
Provide a container algorithm similar to C++ Library Fundamentals v2's erase_if which is equivalent t...
LLVM_ABI bool UpgradeDebugInfo(Module &M)
Check the debug info version number, if it is out-dated, drop the debug info.
LLVM_ABI void UpgradeFunctionAttributes(Function &F)
Correct any IR that is relying on old function attribute behavior.
LLVM_ABI MDNode * UpgradeTBAANode(MDNode &TBAANode)
If the given TBAA tag uses the scalar TBAA format, create a new node corresponding to the upgrade to ...
LLVM_ABI void UpgradeARCRuntime(Module &M)
Convert calls to ARC runtime functions to intrinsic calls and upgrade the old retain release marker t...
@ Default
The result value is uniform if and only if all operands are uniform.
LLVM_ABI bool verifyModule(const Module &M, raw_ostream *OS=nullptr, bool *BrokenDebugInfo=nullptr)
Check a module for errors.
LLVM_ABI void reportFatalUsageError(Error Err)
Report a fatal error that does not indicate a bug in LLVM.
void swap(llvm::BitVector &LHS, llvm::BitVector &RHS)
Implement std::swap in terms of BitVector swap.
This struct is a compact representation of a valid (non-zero power of two) alignment.
Represents the full denormal controls for a function, including the default mode and the f32 specific...
Represent subnormal handling kind for floating point instruction inputs and outputs.
static constexpr DenormalMode getInvalid()
constexpr bool isValid() const
static constexpr DenormalMode getIEEE()
This struct is a compact representation of a valid (power of two) or undefined (0) alignment.