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());
1451 if (Name ==
"coro.end" &&
1452 (
F->arg_size() == 2 ||
F->getReturnType()->isIntegerTy(1)))
1453 CoroEndID = Intrinsic::coro_end;
1454 else if (Name ==
"coro.end.async" &&
F->getReturnType()->isIntegerTy(1))
1455 CoroEndID = Intrinsic::coro_end_async;
1466 if (Name.consume_front(
"dbg.")) {
1468 if (CanUpgradeDebugIntrinsicsToRecords) {
1469 if (Name ==
"addr" || Name ==
"value" || Name ==
"assign" ||
1470 Name ==
"declare" || Name ==
"label") {
1479 if (Name ==
"addr" || (Name ==
"value" &&
F->arg_size() == 4)) {
1482 Intrinsic::dbg_value);
1489 if (Name.consume_front(
"experimental.vector.")) {
1495 .
StartsWith(
"extract.", Intrinsic::vector_extract)
1496 .
StartsWith(
"insert.", Intrinsic::vector_insert)
1497 .
StartsWith(
"reverse.", Intrinsic::vector_reverse)
1498 .
StartsWith(
"interleave2.", Intrinsic::vector_interleave2)
1499 .
StartsWith(
"deinterleave2.", Intrinsic::vector_deinterleave2)
1501 Intrinsic::vector_partial_reduce_add)
1504 const auto *FT =
F->getFunctionType();
1506 if (
ID == Intrinsic::vector_extract ||
1507 ID == Intrinsic::vector_interleave2)
1510 if (
ID != Intrinsic::vector_interleave2)
1512 if (
ID == Intrinsic::vector_insert ||
1513 ID == Intrinsic::vector_partial_reduce_add)
1521 if (Name.consume_front(
"reduce.")) {
1523 static const Regex R(
"^([a-z]+)\\.[a-z][0-9]+");
1524 if (R.match(Name, &
Groups))
1526 .
Case(
"add", Intrinsic::vector_reduce_add)
1527 .
Case(
"mul", Intrinsic::vector_reduce_mul)
1528 .
Case(
"and", Intrinsic::vector_reduce_and)
1529 .
Case(
"or", Intrinsic::vector_reduce_or)
1530 .
Case(
"xor", Intrinsic::vector_reduce_xor)
1531 .
Case(
"smax", Intrinsic::vector_reduce_smax)
1532 .
Case(
"smin", Intrinsic::vector_reduce_smin)
1533 .
Case(
"umax", Intrinsic::vector_reduce_umax)
1534 .
Case(
"umin", Intrinsic::vector_reduce_umin)
1535 .
Case(
"fmax", Intrinsic::vector_reduce_fmax)
1536 .
Case(
"fmin", Intrinsic::vector_reduce_fmin)
1541 static const Regex R2(
"^v2\\.([a-z]+)\\.[fi][0-9]+");
1546 .
Case(
"fadd", Intrinsic::vector_reduce_fadd)
1547 .
Case(
"fmul", Intrinsic::vector_reduce_fmul)
1552 auto Args =
F->getFunctionType()->params();
1554 {Args[V2 ? 1 : 0]});
1560 if (Name.consume_front(
"splice"))
1564 if (Name.consume_front(
"experimental.stepvector.")) {
1568 F->getParent(),
ID,
F->getFunctionType()->getReturnType());
1573 if (Name.starts_with(
"flt.rounds")) {
1576 Intrinsic::get_rounding);
1581 if (Name.starts_with(
"invariant.group.barrier")) {
1583 auto Args =
F->getFunctionType()->params();
1584 Type* ObjectPtr[1] = {Args[0]};
1587 F->getParent(), Intrinsic::launder_invariant_group, ObjectPtr);
1592 bool IsLifetimeStart = Name.consume_front(
"lifetime.start");
1593 bool IsLifetimeEnd = !IsLifetimeStart && Name.consume_front(
"lifetime.end");
1594 if (IsLifetimeStart || IsLifetimeEnd) {
1595 if (
F->arg_size() == 2) {
1596 Intrinsic::ID IID = IsLifetimeStart ? Intrinsic::lifetime_start
1597 : Intrinsic::lifetime_end;
1602 F->getArg(1)->getType());
1604 }
else if (
F->arg_size() == 1 && Name ==
".i64") {
1624 .StartsWith(
"memcpy.", Intrinsic::memcpy)
1625 .StartsWith(
"memmove.", Intrinsic::memmove)
1627 if (
F->arg_size() == 5) {
1631 F->getFunctionType()->params().slice(0, 3);
1637 if (Name.starts_with(
"memset.") &&
F->arg_size() == 5) {
1640 const auto *FT =
F->getFunctionType();
1641 Type *ParamTypes[2] = {
1642 FT->getParamType(0),
1646 Intrinsic::memset, ParamTypes);
1652 .
StartsWith(
"masked.load", Intrinsic::masked_load)
1653 .
StartsWith(
"masked.gather", Intrinsic::masked_gather)
1654 .
StartsWith(
"masked.store", Intrinsic::masked_store)
1655 .
StartsWith(
"masked.scatter", Intrinsic::masked_scatter)
1657 if (MaskedID &&
F->arg_size() == 4) {
1659 if (MaskedID == Intrinsic::masked_load ||
1660 MaskedID == Intrinsic::masked_gather) {
1662 F->getParent(), MaskedID,
1663 {F->getReturnType(), F->getArg(0)->getType()});
1667 F->getParent(), MaskedID,
1668 {F->getArg(0)->getType(), F->getArg(1)->getType()});
1674 if (Name.consume_front(
"nvvm.")) {
1676 if (
F->arg_size() == 1) {
1679 .
Cases({
"brev32",
"brev64"}, Intrinsic::bitreverse)
1680 .Case(
"clz.i", Intrinsic::ctlz)
1681 .
Case(
"popc.i", Intrinsic::ctpop)
1685 {F->getReturnType()});
1688 }
else if (
F->arg_size() == 2) {
1691 .
Cases({
"max.s",
"max.i",
"max.ll"}, Intrinsic::smax)
1692 .Cases({
"min.s",
"min.i",
"min.ll"}, Intrinsic::smin)
1693 .Cases({
"max.us",
"max.ui",
"max.ull"}, Intrinsic::umax)
1694 .Cases({
"min.us",
"min.ui",
"min.ull"}, Intrinsic::umin)
1698 {F->getReturnType()});
1704 if (!
F->getReturnType()->getScalarType()->isBFloatTy()) {
1732 bool Expand =
false;
1733 if (Name.consume_front(
"abs."))
1736 Name ==
"i" || Name ==
"ll" || Name ==
"bf16" || Name ==
"bf16x2";
1737 else if (Name.consume_front(
"fabs."))
1739 Expand = Name ==
"f" || Name ==
"ftz.f" || Name ==
"d";
1740 else if (Name.consume_front(
"ex2.approx."))
1743 Name ==
"f" || Name ==
"ftz.f" || Name ==
"d" || Name ==
"f16x2";
1744 else if (Name.consume_front(
"atomic.load."))
1753 else if (Name.consume_front(
"atomic."))
1768 else if (Name.consume_front(
"bitcast."))
1771 Name ==
"f2i" || Name ==
"i2f" || Name ==
"ll2d" || Name ==
"d2ll";
1772 else if (Name.consume_front(
"rotate."))
1774 Expand = Name ==
"b32" || Name ==
"b64" || Name ==
"right.b64";
1775 else if (Name.consume_front(
"ptr.gen.to."))
1778 else if (Name.consume_front(
"ptr."))
1781 else if (Name.consume_front(
"ldg.global."))
1783 Expand = (Name.starts_with(
"i.") || Name.starts_with(
"f.") ||
1784 Name.starts_with(
"p."));
1787 .
Case(
"barrier0",
true)
1788 .
Case(
"barrier.n",
true)
1789 .
Case(
"barrier.sync.cnt",
true)
1790 .
Case(
"barrier.sync",
true)
1791 .
Case(
"barrier",
true)
1792 .
Case(
"bar.sync",
true)
1793 .
Case(
"barrier0.popc",
true)
1794 .
Case(
"barrier0.and",
true)
1795 .
Case(
"barrier0.or",
true)
1796 .
Case(
"clz.ll",
true)
1797 .
Case(
"popc.ll",
true)
1799 .
Case(
"swap.lo.hi.b64",
true)
1800 .
Case(
"tanh.approx.f32",
true)
1812 if (Name.starts_with(
"objectsize.")) {
1813 Type *Tys[2] = {
F->getReturnType(),
F->arg_begin()->getType() };
1814 if (
F->arg_size() == 2 ||
F->arg_size() == 3) {
1817 Intrinsic::objectsize, Tys);
1824 if (Name.starts_with(
"ptr.annotation.") &&
F->arg_size() == 4) {
1827 F->getParent(), Intrinsic::ptr_annotation,
1828 {F->arg_begin()->getType(), F->getArg(1)->getType()});
1834 if (Name.consume_front(
"riscv.")) {
1837 .
Case(
"aes32dsi", Intrinsic::riscv_aes32dsi)
1838 .
Case(
"aes32dsmi", Intrinsic::riscv_aes32dsmi)
1839 .
Case(
"aes32esi", Intrinsic::riscv_aes32esi)
1840 .
Case(
"aes32esmi", Intrinsic::riscv_aes32esmi)
1843 if (!
F->getFunctionType()->getParamType(2)->isIntegerTy(32)) {
1856 if (!
F->getFunctionType()->getParamType(2)->isIntegerTy(32) ||
1857 F->getFunctionType()->getReturnType()->isIntegerTy(64)) {
1866 .
StartsWith(
"sha256sig0", Intrinsic::riscv_sha256sig0)
1867 .
StartsWith(
"sha256sig1", Intrinsic::riscv_sha256sig1)
1868 .
StartsWith(
"sha256sum0", Intrinsic::riscv_sha256sum0)
1869 .
StartsWith(
"sha256sum1", Intrinsic::riscv_sha256sum1)
1874 if (
F->getFunctionType()->getReturnType()->isIntegerTy(64)) {
1883 if (Name ==
"clmul.i32" || Name ==
"clmul.i64") {
1885 F->getParent(), Intrinsic::clmul, {F->getReturnType()});
1894 if (Name ==
"stackprotectorcheck") {
1901 if (Name ==
"thread.pointer") {
1903 F->getParent(), Intrinsic::thread_pointer,
F->getReturnType());
1909 if (Name ==
"var.annotation" &&
F->arg_size() == 4) {
1912 F->getParent(), Intrinsic::var_annotation,
1913 {{F->arg_begin()->getType(), F->getArg(1)->getType()}});
1916 if (Name.consume_front(
"vector.splice")) {
1917 if (Name.starts_with(
".left") || Name.starts_with(
".right"))
1925 if (Name.consume_front(
"wasm.")) {
1928 .
StartsWith(
"fma.", Intrinsic::wasm_relaxed_madd)
1929 .
StartsWith(
"fms.", Intrinsic::wasm_relaxed_nmadd)
1930 .
StartsWith(
"laneselect.", Intrinsic::wasm_relaxed_laneselect)
1935 F->getReturnType());
1939 if (Name.consume_front(
"dot.i8x16.i7x16.")) {
1941 .
Case(
"signed", Intrinsic::wasm_relaxed_dot_i8x16_i7x16_signed)
1943 Intrinsic::wasm_relaxed_dot_i8x16_i7x16_add_signed)
1962 if (ST && (!
ST->isLiteral() ||
ST->isPacked()) &&
1972 std::string
Name =
F->getName().str();
1975 Name,
F->getParent());
1986 if (Result != std::nullopt) {
2002 bool CanUpgradeDebugIntrinsicsToRecords) {
2022 GV->
getName() ==
"llvm.global_dtors")) ||
2037 unsigned N =
Init->getNumOperands();
2038 std::vector<Constant *> NewCtors(
N);
2039 for (
unsigned i = 0; i !=
N; ++i) {
2042 Ctor->getAggregateElement(1),
2056 unsigned NumElts = ResultTy->getNumElements() * 8;
2060 Op = Builder.CreateBitCast(
Op, VecTy,
"cast");
2070 for (
unsigned l = 0; l != NumElts; l += 16)
2071 for (
unsigned i = 0; i != 16; ++i) {
2072 unsigned Idx = NumElts + i - Shift;
2074 Idx -= NumElts - 16;
2075 Idxs[l + i] = Idx + l;
2078 Res = Builder.CreateShuffleVector(Res,
Op,
ArrayRef(Idxs, NumElts));
2082 return Builder.CreateBitCast(Res, ResultTy,
"cast");
2090 unsigned NumElts = ResultTy->getNumElements() * 8;
2094 Op = Builder.CreateBitCast(
Op, VecTy,
"cast");
2104 for (
unsigned l = 0; l != NumElts; l += 16)
2105 for (
unsigned i = 0; i != 16; ++i) {
2106 unsigned Idx = i + Shift;
2108 Idx += NumElts - 16;
2109 Idxs[l + i] = Idx + l;
2112 Res = Builder.CreateShuffleVector(
Op, Res,
ArrayRef(Idxs, NumElts));
2116 return Builder.CreateBitCast(Res, ResultTy,
"cast");
2124 Mask = Builder.CreateBitCast(Mask, MaskTy);
2130 for (
unsigned i = 0; i != NumElts; ++i)
2132 Mask = Builder.CreateShuffleVector(Mask, Mask,
ArrayRef(Indices, NumElts),
2143 if (
C->isAllOnesValue())
2148 return Builder.CreateSelect(Mask, Op0, Op1);
2155 if (
C->isAllOnesValue())
2159 Mask->getType()->getIntegerBitWidth());
2160 Mask = Builder.CreateBitCast(Mask, MaskTy);
2161 Mask = Builder.CreateExtractElement(Mask, (
uint64_t)0);
2162 return Builder.CreateSelect(Mask, Op0, Op1);
2175 assert((IsVALIGN || NumElts % 16 == 0) &&
"Illegal NumElts for PALIGNR!");
2176 assert((!IsVALIGN || NumElts <= 16) &&
"NumElts too large for VALIGN!");
2181 ShiftVal &= (NumElts - 1);
2190 if (ShiftVal > 16) {
2198 for (
unsigned l = 0; l < NumElts; l += 16) {
2199 for (
unsigned i = 0; i != 16; ++i) {
2200 unsigned Idx = ShiftVal + i;
2201 if (!IsVALIGN && Idx >= 16)
2202 Idx += NumElts - 16;
2203 Indices[l + i] = Idx + l;
2208 Op1, Op0,
ArrayRef(Indices, NumElts),
"palignr");
2214 bool ZeroMask,
bool IndexForm) {
2217 unsigned EltWidth = Ty->getScalarSizeInBits();
2218 bool IsFloat = Ty->isFPOrFPVectorTy();
2220 if (VecWidth == 128 && EltWidth == 32 && IsFloat)
2221 IID = Intrinsic::x86_avx512_vpermi2var_ps_128;
2222 else if (VecWidth == 128 && EltWidth == 32 && !IsFloat)
2223 IID = Intrinsic::x86_avx512_vpermi2var_d_128;
2224 else if (VecWidth == 128 && EltWidth == 64 && IsFloat)
2225 IID = Intrinsic::x86_avx512_vpermi2var_pd_128;
2226 else if (VecWidth == 128 && EltWidth == 64 && !IsFloat)
2227 IID = Intrinsic::x86_avx512_vpermi2var_q_128;
2228 else if (VecWidth == 256 && EltWidth == 32 && IsFloat)
2229 IID = Intrinsic::x86_avx512_vpermi2var_ps_256;
2230 else if (VecWidth == 256 && EltWidth == 32 && !IsFloat)
2231 IID = Intrinsic::x86_avx512_vpermi2var_d_256;
2232 else if (VecWidth == 256 && EltWidth == 64 && IsFloat)
2233 IID = Intrinsic::x86_avx512_vpermi2var_pd_256;
2234 else if (VecWidth == 256 && EltWidth == 64 && !IsFloat)
2235 IID = Intrinsic::x86_avx512_vpermi2var_q_256;
2236 else if (VecWidth == 512 && EltWidth == 32 && IsFloat)
2237 IID = Intrinsic::x86_avx512_vpermi2var_ps_512;
2238 else if (VecWidth == 512 && EltWidth == 32 && !IsFloat)
2239 IID = Intrinsic::x86_avx512_vpermi2var_d_512;
2240 else if (VecWidth == 512 && EltWidth == 64 && IsFloat)
2241 IID = Intrinsic::x86_avx512_vpermi2var_pd_512;
2242 else if (VecWidth == 512 && EltWidth == 64 && !IsFloat)
2243 IID = Intrinsic::x86_avx512_vpermi2var_q_512;
2244 else if (VecWidth == 128 && EltWidth == 16)
2245 IID = Intrinsic::x86_avx512_vpermi2var_hi_128;
2246 else if (VecWidth == 256 && EltWidth == 16)
2247 IID = Intrinsic::x86_avx512_vpermi2var_hi_256;
2248 else if (VecWidth == 512 && EltWidth == 16)
2249 IID = Intrinsic::x86_avx512_vpermi2var_hi_512;
2250 else if (VecWidth == 128 && EltWidth == 8)
2251 IID = Intrinsic::x86_avx512_vpermi2var_qi_128;
2252 else if (VecWidth == 256 && EltWidth == 8)
2253 IID = Intrinsic::x86_avx512_vpermi2var_qi_256;
2254 else if (VecWidth == 512 && EltWidth == 8)
2255 IID = Intrinsic::x86_avx512_vpermi2var_qi_512;
2266 Value *V = Builder.CreateIntrinsic(IID, Args);
2278 Value *Res = Builder.CreateIntrinsic(IID, Ty, {Op0, Op1});
2289 bool IsRotateRight) {
2299 Amt = Builder.CreateIntCast(Amt, Ty->getScalarType(),
false);
2300 Amt = Builder.CreateVectorSplat(NumElts, Amt);
2303 Intrinsic::ID IID = IsRotateRight ? Intrinsic::fshr : Intrinsic::fshl;
2304 Value *Res = Builder.CreateIntrinsic(IID, Ty, {Src, Src, Amt});
2349 Value *Ext = Builder.CreateSExt(Cmp, Ty);
2354 bool IsShiftRight,
bool ZeroMask) {
2368 Amt = Builder.CreateIntCast(Amt, Ty->getScalarType(),
false);
2369 Amt = Builder.CreateVectorSplat(NumElts, Amt);
2372 Intrinsic::ID IID = IsShiftRight ? Intrinsic::fshr : Intrinsic::fshl;
2373 Value *Res = Builder.CreateIntrinsic(IID, Ty, {Op0, Op1, Amt});
2388 const Align Alignment =
2390 ?
Align(
Data->getType()->getPrimitiveSizeInBits().getFixedValue() / 8)
2395 if (
C->isAllOnesValue())
2396 return Builder.CreateAlignedStore(
Data, Ptr, Alignment);
2401 return Builder.CreateMaskedStore(
Data, Ptr, Alignment, Mask);
2407 const Align Alignment =
2416 if (
C->isAllOnesValue())
2417 return Builder.CreateAlignedLoad(ValTy, Ptr, Alignment);
2422 return Builder.CreateMaskedLoad(ValTy, Ptr, Alignment, Mask, Passthru);
2428 Value *Res = Builder.CreateIntrinsic(Intrinsic::abs, Ty,
2429 {Op0, Builder.getInt1(
false)});
2444 Constant *ShiftAmt = ConstantInt::get(Ty, 32);
2445 LHS = Builder.CreateShl(
LHS, ShiftAmt);
2446 LHS = Builder.CreateAShr(
LHS, ShiftAmt);
2447 RHS = Builder.CreateShl(
RHS, ShiftAmt);
2448 RHS = Builder.CreateAShr(
RHS, ShiftAmt);
2451 Constant *Mask = ConstantInt::get(Ty, 0xffffffff);
2452 LHS = Builder.CreateAnd(
LHS, Mask);
2453 RHS = Builder.CreateAnd(
RHS, Mask);
2470 if (!
C || !
C->isAllOnesValue())
2471 Vec = Builder.CreateAnd(Vec,
getX86MaskVec(Builder, Mask, NumElts));
2476 for (
unsigned i = 0; i != NumElts; ++i)
2478 for (
unsigned i = NumElts; i != 8; ++i)
2479 Indices[i] = NumElts + i % NumElts;
2480 Vec = Builder.CreateShuffleVector(Vec,
2484 return Builder.CreateBitCast(Vec, Builder.getIntNTy(std::max(NumElts, 8U)));
2488 unsigned CC,
bool Signed) {
2496 }
else if (CC == 7) {
2532 Value* AndNode = Builder.CreateAnd(Mask,
APInt(8, 1));
2533 Value* Cmp = Builder.CreateIsNotNull(AndNode);
2535 Value* Extract2 = Builder.CreateExtractElement(Src, (
uint64_t)0);
2536 Value*
Select = Builder.CreateSelect(Cmp, Extract1, Extract2);
2545 return Builder.CreateSExt(Mask, ReturnOp,
"vpmovm2");
2551 Name = Name.substr(12);
2556 if (Name.starts_with(
"max.p")) {
2557 if (VecWidth == 128 && EltWidth == 32)
2558 IID = Intrinsic::x86_sse_max_ps;
2559 else if (VecWidth == 128 && EltWidth == 64)
2560 IID = Intrinsic::x86_sse2_max_pd;
2561 else if (VecWidth == 256 && EltWidth == 32)
2562 IID = Intrinsic::x86_avx_max_ps_256;
2563 else if (VecWidth == 256 && EltWidth == 64)
2564 IID = Intrinsic::x86_avx_max_pd_256;
2567 }
else if (Name.starts_with(
"min.p")) {
2568 if (VecWidth == 128 && EltWidth == 32)
2569 IID = Intrinsic::x86_sse_min_ps;
2570 else if (VecWidth == 128 && EltWidth == 64)
2571 IID = Intrinsic::x86_sse2_min_pd;
2572 else if (VecWidth == 256 && EltWidth == 32)
2573 IID = Intrinsic::x86_avx_min_ps_256;
2574 else if (VecWidth == 256 && EltWidth == 64)
2575 IID = Intrinsic::x86_avx_min_pd_256;
2578 }
else if (Name.starts_with(
"pshuf.b.")) {
2579 if (VecWidth == 128)
2580 IID = Intrinsic::x86_ssse3_pshuf_b_128;
2581 else if (VecWidth == 256)
2582 IID = Intrinsic::x86_avx2_pshuf_b;
2583 else if (VecWidth == 512)
2584 IID = Intrinsic::x86_avx512_pshuf_b_512;
2587 }
else if (Name.starts_with(
"pmul.hr.sw.")) {
2588 if (VecWidth == 128)
2589 IID = Intrinsic::x86_ssse3_pmul_hr_sw_128;
2590 else if (VecWidth == 256)
2591 IID = Intrinsic::x86_avx2_pmul_hr_sw;
2592 else if (VecWidth == 512)
2593 IID = Intrinsic::x86_avx512_pmul_hr_sw_512;
2596 }
else if (Name.starts_with(
"pmulh.w.")) {
2597 if (VecWidth == 128)
2598 IID = Intrinsic::x86_sse2_pmulh_w;
2599 else if (VecWidth == 256)
2600 IID = Intrinsic::x86_avx2_pmulh_w;
2601 else if (VecWidth == 512)
2602 IID = Intrinsic::x86_avx512_pmulh_w_512;
2605 }
else if (Name.starts_with(
"pmulhu.w.")) {
2606 if (VecWidth == 128)
2607 IID = Intrinsic::x86_sse2_pmulhu_w;
2608 else if (VecWidth == 256)
2609 IID = Intrinsic::x86_avx2_pmulhu_w;
2610 else if (VecWidth == 512)
2611 IID = Intrinsic::x86_avx512_pmulhu_w_512;
2614 }
else if (Name.starts_with(
"pmaddw.d.")) {
2615 if (VecWidth == 128)
2616 IID = Intrinsic::x86_sse2_pmadd_wd;
2617 else if (VecWidth == 256)
2618 IID = Intrinsic::x86_avx2_pmadd_wd;
2619 else if (VecWidth == 512)
2620 IID = Intrinsic::x86_avx512_pmaddw_d_512;
2623 }
else if (Name.starts_with(
"pmaddubs.w.")) {
2624 if (VecWidth == 128)
2625 IID = Intrinsic::x86_ssse3_pmadd_ub_sw_128;
2626 else if (VecWidth == 256)
2627 IID = Intrinsic::x86_avx2_pmadd_ub_sw;
2628 else if (VecWidth == 512)
2629 IID = Intrinsic::x86_avx512_pmaddubs_w_512;
2632 }
else if (Name.starts_with(
"packsswb.")) {
2633 if (VecWidth == 128)
2634 IID = Intrinsic::x86_sse2_packsswb_128;
2635 else if (VecWidth == 256)
2636 IID = Intrinsic::x86_avx2_packsswb;
2637 else if (VecWidth == 512)
2638 IID = Intrinsic::x86_avx512_packsswb_512;
2641 }
else if (Name.starts_with(
"packssdw.")) {
2642 if (VecWidth == 128)
2643 IID = Intrinsic::x86_sse2_packssdw_128;
2644 else if (VecWidth == 256)
2645 IID = Intrinsic::x86_avx2_packssdw;
2646 else if (VecWidth == 512)
2647 IID = Intrinsic::x86_avx512_packssdw_512;
2650 }
else if (Name.starts_with(
"packuswb.")) {
2651 if (VecWidth == 128)
2652 IID = Intrinsic::x86_sse2_packuswb_128;
2653 else if (VecWidth == 256)
2654 IID = Intrinsic::x86_avx2_packuswb;
2655 else if (VecWidth == 512)
2656 IID = Intrinsic::x86_avx512_packuswb_512;
2659 }
else if (Name.starts_with(
"packusdw.")) {
2660 if (VecWidth == 128)
2661 IID = Intrinsic::x86_sse41_packusdw;
2662 else if (VecWidth == 256)
2663 IID = Intrinsic::x86_avx2_packusdw;
2664 else if (VecWidth == 512)
2665 IID = Intrinsic::x86_avx512_packusdw_512;
2668 }
else if (Name.starts_with(
"vpermilvar.")) {
2669 if (VecWidth == 128 && EltWidth == 32)
2670 IID = Intrinsic::x86_avx_vpermilvar_ps;
2671 else if (VecWidth == 128 && EltWidth == 64)
2672 IID = Intrinsic::x86_avx_vpermilvar_pd;
2673 else if (VecWidth == 256 && EltWidth == 32)
2674 IID = Intrinsic::x86_avx_vpermilvar_ps_256;
2675 else if (VecWidth == 256 && EltWidth == 64)
2676 IID = Intrinsic::x86_avx_vpermilvar_pd_256;
2677 else if (VecWidth == 512 && EltWidth == 32)
2678 IID = Intrinsic::x86_avx512_vpermilvar_ps_512;
2679 else if (VecWidth == 512 && EltWidth == 64)
2680 IID = Intrinsic::x86_avx512_vpermilvar_pd_512;
2683 }
else if (Name ==
"cvtpd2dq.256") {
2684 IID = Intrinsic::x86_avx_cvt_pd2dq_256;
2685 }
else if (Name ==
"cvtpd2ps.256") {
2686 IID = Intrinsic::x86_avx_cvt_pd2_ps_256;
2687 }
else if (Name ==
"cvttpd2dq.256") {
2688 IID = Intrinsic::x86_avx_cvtt_pd2dq_256;
2689 }
else if (Name ==
"cvttps2dq.128") {
2690 IID = Intrinsic::x86_sse2_cvttps2dq;
2691 }
else if (Name ==
"cvttps2dq.256") {
2692 IID = Intrinsic::x86_avx_cvtt_ps2dq_256;
2693 }
else if (Name.starts_with(
"permvar.")) {
2695 if (VecWidth == 256 && EltWidth == 32 && IsFloat)
2696 IID = Intrinsic::x86_avx2_permps;
2697 else if (VecWidth == 256 && EltWidth == 32 && !IsFloat)
2698 IID = Intrinsic::x86_avx2_permd;
2699 else if (VecWidth == 256 && EltWidth == 64 && IsFloat)
2700 IID = Intrinsic::x86_avx512_permvar_df_256;
2701 else if (VecWidth == 256 && EltWidth == 64 && !IsFloat)
2702 IID = Intrinsic::x86_avx512_permvar_di_256;
2703 else if (VecWidth == 512 && EltWidth == 32 && IsFloat)
2704 IID = Intrinsic::x86_avx512_permvar_sf_512;
2705 else if (VecWidth == 512 && EltWidth == 32 && !IsFloat)
2706 IID = Intrinsic::x86_avx512_permvar_si_512;
2707 else if (VecWidth == 512 && EltWidth == 64 && IsFloat)
2708 IID = Intrinsic::x86_avx512_permvar_df_512;
2709 else if (VecWidth == 512 && EltWidth == 64 && !IsFloat)
2710 IID = Intrinsic::x86_avx512_permvar_di_512;
2711 else if (VecWidth == 128 && EltWidth == 16)
2712 IID = Intrinsic::x86_avx512_permvar_hi_128;
2713 else if (VecWidth == 256 && EltWidth == 16)
2714 IID = Intrinsic::x86_avx512_permvar_hi_256;
2715 else if (VecWidth == 512 && EltWidth == 16)
2716 IID = Intrinsic::x86_avx512_permvar_hi_512;
2717 else if (VecWidth == 128 && EltWidth == 8)
2718 IID = Intrinsic::x86_avx512_permvar_qi_128;
2719 else if (VecWidth == 256 && EltWidth == 8)
2720 IID = Intrinsic::x86_avx512_permvar_qi_256;
2721 else if (VecWidth == 512 && EltWidth == 8)
2722 IID = Intrinsic::x86_avx512_permvar_qi_512;
2725 }
else if (Name.starts_with(
"dbpsadbw.")) {
2726 if (VecWidth == 128)
2727 IID = Intrinsic::x86_avx512_dbpsadbw_128;
2728 else if (VecWidth == 256)
2729 IID = Intrinsic::x86_avx512_dbpsadbw_256;
2730 else if (VecWidth == 512)
2731 IID = Intrinsic::x86_avx512_dbpsadbw_512;
2734 }
else if (Name.starts_with(
"pmultishift.qb.")) {
2735 if (VecWidth == 128)
2736 IID = Intrinsic::x86_avx512_pmultishift_qb_128;
2737 else if (VecWidth == 256)
2738 IID = Intrinsic::x86_avx512_pmultishift_qb_256;
2739 else if (VecWidth == 512)
2740 IID = Intrinsic::x86_avx512_pmultishift_qb_512;
2743 }
else if (Name.starts_with(
"conflict.")) {
2744 if (Name[9] ==
'd' && VecWidth == 128)
2745 IID = Intrinsic::x86_avx512_conflict_d_128;
2746 else if (Name[9] ==
'd' && VecWidth == 256)
2747 IID = Intrinsic::x86_avx512_conflict_d_256;
2748 else if (Name[9] ==
'd' && VecWidth == 512)
2749 IID = Intrinsic::x86_avx512_conflict_d_512;
2750 else if (Name[9] ==
'q' && VecWidth == 128)
2751 IID = Intrinsic::x86_avx512_conflict_q_128;
2752 else if (Name[9] ==
'q' && VecWidth == 256)
2753 IID = Intrinsic::x86_avx512_conflict_q_256;
2754 else if (Name[9] ==
'q' && VecWidth == 512)
2755 IID = Intrinsic::x86_avx512_conflict_q_512;
2758 }
else if (Name.starts_with(
"pavg.")) {
2759 if (Name[5] ==
'b' && VecWidth == 128)
2760 IID = Intrinsic::x86_sse2_pavg_b;
2761 else if (Name[5] ==
'b' && VecWidth == 256)
2762 IID = Intrinsic::x86_avx2_pavg_b;
2763 else if (Name[5] ==
'b' && VecWidth == 512)
2764 IID = Intrinsic::x86_avx512_pavg_b_512;
2765 else if (Name[5] ==
'w' && VecWidth == 128)
2766 IID = Intrinsic::x86_sse2_pavg_w;
2767 else if (Name[5] ==
'w' && VecWidth == 256)
2768 IID = Intrinsic::x86_avx2_pavg_w;
2769 else if (Name[5] ==
'w' && VecWidth == 512)
2770 IID = Intrinsic::x86_avx512_pavg_w_512;
2779 Rep = Builder.CreateIntrinsic(IID, Args);
2790 if (AsmStr->find(
"mov\tfp") == 0 &&
2791 AsmStr->find(
"objc_retainAutoreleaseReturnValue") != std::string::npos &&
2792 (Pos = AsmStr->find(
"# marker")) != std::string::npos) {
2793 AsmStr->replace(Pos, 1,
";");
2799 Value *Rep =
nullptr;
2801 if (Name ==
"abs.i" || Name ==
"abs.ll") {
2803 Rep = Builder.CreateIntrinsic(Intrinsic::abs, {Arg->
getType()},
2804 {Arg, Builder.getTrue()},
2806 }
else if (Name ==
"abs.bf16" || Name ==
"abs.bf16x2") {
2807 Type *Ty = (Name ==
"abs.bf16")
2811 Value *Abs = Builder.CreateUnaryIntrinsic(Intrinsic::nvvm_fabs, Arg);
2812 Rep = Builder.CreateBitCast(Abs, CI->
getType());
2813 }
else if (Name ==
"fabs.f" || Name ==
"fabs.ftz.f" || Name ==
"fabs.d") {
2814 Intrinsic::ID IID = (Name ==
"fabs.ftz.f") ? Intrinsic::nvvm_fabs_ftz
2815 : Intrinsic::nvvm_fabs;
2816 Rep = Builder.CreateUnaryIntrinsic(IID, CI->
getArgOperand(0));
2817 }
else if (Name.consume_front(
"ex2.approx.")) {
2819 Intrinsic::ID IID = Name.starts_with(
"ftz") ? Intrinsic::nvvm_ex2_approx_ftz
2820 : Intrinsic::nvvm_ex2_approx;
2821 Rep = Builder.CreateUnaryIntrinsic(IID, CI->
getArgOperand(0));
2822 }
else if (Name.starts_with(
"atomic.load.add.f32.p") ||
2823 Name.starts_with(
"atomic.load.add.f64.p")) {
2826 Rep = Builder.CreateAtomicRMW(
2832 }
else if (Name.starts_with(
"atomic.load.inc.32.p") ||
2833 Name.starts_with(
"atomic.load.dec.32.p")) {
2838 Rep = Builder.CreateAtomicRMW(
2842 }
else if (Name.starts_with(
"atomic.") && Name.contains(
".gen.")) {
2848 Op.contains(
".cta.") ?
"block" :
"");
2849 if (
Op.starts_with(
"cas.")) {
2851 Value *Pair = Builder.CreateAtomicCmpXchg(
2854 Rep = Builder.CreateExtractValue(Pair, 0);
2872 "unexpected nvvm scoped atomic intrinsic");
2873 Rep = Builder.CreateAtomicRMW(BinOp, Ptr, Val,
MaybeAlign(),
2876 }
else if (Name ==
"clz.ll") {
2879 Value *Ctlz = Builder.CreateIntrinsic(Intrinsic::ctlz, {Arg->
getType()},
2880 {Arg, Builder.getFalse()},
2882 Rep = Builder.CreateTrunc(Ctlz, Builder.getInt32Ty(),
"ctlz.trunc");
2883 }
else if (Name ==
"popc.ll") {
2887 Value *Popc = Builder.CreateIntrinsic(Intrinsic::ctpop, {Arg->
getType()},
2888 Arg,
nullptr,
"ctpop");
2889 Rep = Builder.CreateTrunc(Popc, Builder.getInt32Ty(),
"ctpop.trunc");
2890 }
else if (Name ==
"h2f") {
2892 Builder.CreateBitCast(CI->
getArgOperand(0), Builder.getHalfTy());
2893 Rep = Builder.CreateFPExt(Cast, Builder.getFloatTy());
2894 }
else if (Name.consume_front(
"bitcast.") &&
2895 (Name ==
"f2i" || Name ==
"i2f" || Name ==
"ll2d" ||
2898 }
else if (Name ==
"rotate.b32") {
2901 Rep = Builder.CreateIntrinsic(Builder.getInt32Ty(), Intrinsic::fshl,
2902 {Arg, Arg, ShiftAmt});
2903 }
else if (Name ==
"rotate.b64") {
2907 Rep = Builder.CreateIntrinsic(Int64Ty, Intrinsic::fshl,
2908 {Arg, Arg, ZExtShiftAmt});
2909 }
else if (Name ==
"rotate.right.b64") {
2913 Rep = Builder.CreateIntrinsic(Int64Ty, Intrinsic::fshr,
2914 {Arg, Arg, ZExtShiftAmt});
2915 }
else if (Name ==
"swap.lo.hi.b64") {
2918 Rep = Builder.CreateIntrinsic(Int64Ty, Intrinsic::fshl,
2919 {Arg, Arg, Builder.getInt64(32)});
2920 }
else if ((Name.consume_front(
"ptr.gen.to.") &&
2923 Name.starts_with(
".to.gen"))) {
2925 }
else if (Name.consume_front(
"ldg.global")) {
2929 Value *ASC = Builder.CreateAddrSpaceCast(Ptr, Builder.getPtrTy(1));
2932 LD->setMetadata(LLVMContext::MD_invariant_load, MD);
2934 }
else if (Name ==
"tanh.approx.f32") {
2938 Rep = Builder.CreateUnaryIntrinsic(Intrinsic::tanh, CI->
getArgOperand(0),
2940 }
else if (Name ==
"barrier0" || Name ==
"barrier.n" || Name ==
"bar.sync") {
2942 Name.ends_with(
'0') ? Builder.getInt32(0) : CI->
getArgOperand(0);
2943 Rep = Builder.CreateIntrinsic(Intrinsic::nvvm_barrier_cta_sync_aligned_all,
2945 }
else if (Name ==
"barrier") {
2946 Rep = Builder.CreateIntrinsic(
2947 Intrinsic::nvvm_barrier_cta_sync_aligned_count, {},
2949 }
else if (Name ==
"barrier.sync") {
2950 Rep = Builder.CreateIntrinsic(Intrinsic::nvvm_barrier_cta_sync_all, {},
2952 }
else if (Name ==
"barrier.sync.cnt") {
2953 Rep = Builder.CreateIntrinsic(Intrinsic::nvvm_barrier_cta_sync_count, {},
2955 }
else if (Name ==
"barrier0.popc" || Name ==
"barrier0.and" ||
2956 Name ==
"barrier0.or") {
2958 C = Builder.CreateICmpNE(
C, Builder.getInt32(0));
2962 .
Case(
"barrier0.popc",
2963 Intrinsic::nvvm_barrier_cta_red_popc_aligned_all)
2964 .
Case(
"barrier0.and",
2965 Intrinsic::nvvm_barrier_cta_red_and_aligned_all)
2966 .
Case(
"barrier0.or",
2967 Intrinsic::nvvm_barrier_cta_red_or_aligned_all);
2968 Value *Bar = Builder.CreateIntrinsic(IID, {}, {Builder.getInt32(0),
C});
2969 Rep = Builder.CreateZExt(Bar, CI->
getType());
2973 !
F->getReturnType()->getScalarType()->isBFloatTy()) {
2983 ? Builder.CreateBitCast(Arg, NewType)
2986 Rep = Builder.CreateCall(NewFn, Args);
2987 if (
F->getReturnType()->isIntegerTy())
2988 Rep = Builder.CreateBitCast(Rep,
F->getReturnType());
2998 Value *Rep =
nullptr;
3000 if (Name.starts_with(
"sse4a.movnt.")) {
3012 Builder.CreateExtractElement(Arg1, (
uint64_t)0,
"extractelement");
3015 SI->setMetadata(LLVMContext::MD_nontemporal,
Node);
3016 }
else if (Name.starts_with(
"avx.movnt.") ||
3017 Name.starts_with(
"avx512.storent.")) {
3029 SI->setMetadata(LLVMContext::MD_nontemporal,
Node);
3030 }
else if (Name ==
"sse2.storel.dq") {
3035 Value *BC0 = Builder.CreateBitCast(Arg1, NewVecTy,
"cast");
3036 Value *Elt = Builder.CreateExtractElement(BC0, (
uint64_t)0);
3037 Builder.CreateAlignedStore(Elt, Arg0,
Align(1));
3038 }
else if (Name.starts_with(
"sse.storeu.") ||
3039 Name.starts_with(
"sse2.storeu.") ||
3040 Name.starts_with(
"avx.storeu.")) {
3043 Builder.CreateAlignedStore(Arg1, Arg0,
Align(1));
3044 }
else if (Name ==
"avx512.mask.store.ss") {
3048 }
else if (Name.starts_with(
"avx512.mask.store")) {
3050 bool Aligned = Name[17] !=
'u';
3053 }
else if (Name.starts_with(
"sse2.pcmp") || Name.starts_with(
"avx2.pcmp")) {
3056 bool CmpEq = Name[9] ==
'e';
3059 Rep = Builder.CreateSExt(Rep, CI->
getType(),
"");
3060 }
else if (Name.starts_with(
"avx512.broadcastm")) {
3067 Rep = Builder.CreateVectorSplat(NumElts, Rep);
3068 }
else if (Name ==
"sse.sqrt.ss" || Name ==
"sse2.sqrt.sd") {
3070 Value *Elt0 = Builder.CreateExtractElement(Vec, (
uint64_t)0);
3071 Elt0 = Builder.CreateIntrinsic(Intrinsic::sqrt, Elt0->
getType(), Elt0);
3072 Rep = Builder.CreateInsertElement(Vec, Elt0, (
uint64_t)0);
3073 }
else if (Name.starts_with(
"avx.sqrt.p") ||
3074 Name.starts_with(
"sse2.sqrt.p") ||
3075 Name.starts_with(
"sse.sqrt.p")) {
3076 Rep = Builder.CreateIntrinsic(Intrinsic::sqrt, CI->
getType(),
3077 {CI->getArgOperand(0)});
3078 }
else if (Name.starts_with(
"avx512.mask.sqrt.p")) {
3082 Intrinsic::ID IID = Name[18] ==
's' ? Intrinsic::x86_avx512_sqrt_ps_512
3083 : Intrinsic::x86_avx512_sqrt_pd_512;
3086 Rep = Builder.CreateIntrinsic(IID, Args);
3088 Rep = Builder.CreateIntrinsic(Intrinsic::sqrt, CI->
getType(),
3089 {CI->getArgOperand(0)});
3093 }
else if (Name.starts_with(
"avx512.ptestm") ||
3094 Name.starts_with(
"avx512.ptestnm")) {
3098 Rep = Builder.CreateAnd(Op0, Op1);
3104 Rep = Builder.CreateICmp(Pred, Rep, Zero);
3106 }
else if (Name.starts_with(
"avx512.mask.pbroadcast")) {
3109 Rep = Builder.CreateVectorSplat(NumElts, CI->
getArgOperand(0));
3112 }
else if (Name.starts_with(
"avx512.kunpck")) {
3117 for (
unsigned i = 0; i != NumElts; ++i)
3126 Rep = Builder.CreateShuffleVector(
RHS,
LHS,
ArrayRef(Indices, NumElts));
3127 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3128 }
else if (Name ==
"avx512.kand.w") {
3131 Rep = Builder.CreateAnd(
LHS,
RHS);
3132 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3133 }
else if (Name ==
"avx512.kandn.w") {
3136 LHS = Builder.CreateNot(
LHS);
3137 Rep = Builder.CreateAnd(
LHS,
RHS);
3138 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3139 }
else if (Name ==
"avx512.kor.w") {
3142 Rep = Builder.CreateOr(
LHS,
RHS);
3143 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3144 }
else if (Name ==
"avx512.kxor.w") {
3147 Rep = Builder.CreateXor(
LHS,
RHS);
3148 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3149 }
else if (Name ==
"avx512.kxnor.w") {
3152 LHS = Builder.CreateNot(
LHS);
3153 Rep = Builder.CreateXor(
LHS,
RHS);
3154 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3155 }
else if (Name ==
"avx512.knot.w") {
3157 Rep = Builder.CreateNot(Rep);
3158 Rep = Builder.CreateBitCast(Rep, CI->
getType());
3159 }
else if (Name ==
"avx512.kortestz.w" || Name ==
"avx512.kortestc.w") {
3162 Rep = Builder.CreateOr(
LHS,
RHS);
3163 Rep = Builder.CreateBitCast(Rep, Builder.getInt16Ty());
3165 if (Name[14] ==
'c')
3169 Rep = Builder.CreateICmpEQ(Rep,
C);
3170 Rep = Builder.CreateZExt(Rep, Builder.getInt32Ty());
3171 }
else if (Name ==
"sse.add.ss" || Name ==
"sse2.add.sd" ||
3172 Name ==
"sse.sub.ss" || Name ==
"sse2.sub.sd" ||
3173 Name ==
"sse.mul.ss" || Name ==
"sse2.mul.sd" ||
3174 Name ==
"sse.div.ss" || Name ==
"sse2.div.sd") {
3177 ConstantInt::get(I32Ty, 0));
3179 ConstantInt::get(I32Ty, 0));
3181 if (Name.contains(
".add."))
3182 EltOp = Builder.CreateFAdd(Elt0, Elt1);
3183 else if (Name.contains(
".sub."))
3184 EltOp = Builder.CreateFSub(Elt0, Elt1);
3185 else if (Name.contains(
".mul."))
3186 EltOp = Builder.CreateFMul(Elt0, Elt1);
3188 EltOp = Builder.CreateFDiv(Elt0, Elt1);
3189 Rep = Builder.CreateInsertElement(CI->
getArgOperand(0), EltOp,
3190 ConstantInt::get(I32Ty, 0));
3191 }
else if (Name.starts_with(
"avx512.mask.pcmp")) {
3193 bool CmpEq = Name[16] ==
'e';
3195 }
else if (Name.starts_with(
"avx512.mask.vpshufbitqmb.")) {
3197 unsigned VecWidth =
OpTy->getPrimitiveSizeInBits();
3204 IID = Intrinsic::x86_avx512_vpshufbitqmb_128;
3207 IID = Intrinsic::x86_avx512_vpshufbitqmb_256;
3210 IID = Intrinsic::x86_avx512_vpshufbitqmb_512;
3217 }
else if (Name.starts_with(
"avx512.mask.fpclass.p")) {
3219 unsigned VecWidth =
OpTy->getPrimitiveSizeInBits();
3220 unsigned EltWidth =
OpTy->getScalarSizeInBits();
3222 if (VecWidth == 128 && EltWidth == 32)
3223 IID = Intrinsic::x86_avx512_fpclass_ps_128;
3224 else if (VecWidth == 256 && EltWidth == 32)
3225 IID = Intrinsic::x86_avx512_fpclass_ps_256;
3226 else if (VecWidth == 512 && EltWidth == 32)
3227 IID = Intrinsic::x86_avx512_fpclass_ps_512;
3228 else if (VecWidth == 128 && EltWidth == 64)
3229 IID = Intrinsic::x86_avx512_fpclass_pd_128;
3230 else if (VecWidth == 256 && EltWidth == 64)
3231 IID = Intrinsic::x86_avx512_fpclass_pd_256;
3232 else if (VecWidth == 512 && EltWidth == 64)
3233 IID = Intrinsic::x86_avx512_fpclass_pd_512;
3240 }
else if (Name.starts_with(
"avx512.cmp.p")) {
3243 unsigned VecWidth =
OpTy->getPrimitiveSizeInBits();
3244 unsigned EltWidth =
OpTy->getScalarSizeInBits();
3246 if (VecWidth == 128 && EltWidth == 32)
3247 IID = Intrinsic::x86_avx512_mask_cmp_ps_128;
3248 else if (VecWidth == 256 && EltWidth == 32)
3249 IID = Intrinsic::x86_avx512_mask_cmp_ps_256;
3250 else if (VecWidth == 512 && EltWidth == 32)
3251 IID = Intrinsic::x86_avx512_mask_cmp_ps_512;
3252 else if (VecWidth == 128 && EltWidth == 64)
3253 IID = Intrinsic::x86_avx512_mask_cmp_pd_128;
3254 else if (VecWidth == 256 && EltWidth == 64)
3255 IID = Intrinsic::x86_avx512_mask_cmp_pd_256;
3256 else if (VecWidth == 512 && EltWidth == 64)
3257 IID = Intrinsic::x86_avx512_mask_cmp_pd_512;
3262 if (VecWidth == 512)
3264 Args.push_back(Mask);
3266 Rep = Builder.CreateIntrinsic(IID, Args);
3267 }
else if (Name.starts_with(
"avx512.mask.cmp.")) {
3271 }
else if (Name.starts_with(
"avx512.mask.ucmp.")) {
3274 }
else if (Name.starts_with(
"avx512.cvtb2mask.") ||
3275 Name.starts_with(
"avx512.cvtw2mask.") ||
3276 Name.starts_with(
"avx512.cvtd2mask.") ||
3277 Name.starts_with(
"avx512.cvtq2mask.")) {
3282 }
else if (Name ==
"ssse3.pabs.b.128" || Name ==
"ssse3.pabs.w.128" ||
3283 Name ==
"ssse3.pabs.d.128" || Name.starts_with(
"avx2.pabs") ||
3284 Name.starts_with(
"avx512.mask.pabs")) {
3286 }
else if (Name ==
"sse41.pmaxsb" || Name ==
"sse2.pmaxs.w" ||
3287 Name ==
"sse41.pmaxsd" || Name.starts_with(
"avx2.pmaxs") ||
3288 Name.starts_with(
"avx512.mask.pmaxs")) {
3290 }
else if (Name ==
"sse2.pmaxu.b" || Name ==
"sse41.pmaxuw" ||
3291 Name ==
"sse41.pmaxud" || Name.starts_with(
"avx2.pmaxu") ||
3292 Name.starts_with(
"avx512.mask.pmaxu")) {
3294 }
else if (Name ==
"sse41.pminsb" || Name ==
"sse2.pmins.w" ||
3295 Name ==
"sse41.pminsd" || Name.starts_with(
"avx2.pmins") ||
3296 Name.starts_with(
"avx512.mask.pmins")) {
3298 }
else if (Name ==
"sse2.pminu.b" || Name ==
"sse41.pminuw" ||
3299 Name ==
"sse41.pminud" || Name.starts_with(
"avx2.pminu") ||
3300 Name.starts_with(
"avx512.mask.pminu")) {
3302 }
else if (Name ==
"sse2.pmulu.dq" || Name ==
"avx2.pmulu.dq" ||
3303 Name ==
"avx512.pmulu.dq.512" ||
3304 Name.starts_with(
"avx512.mask.pmulu.dq.")) {
3306 }
else if (Name ==
"sse41.pmuldq" || Name ==
"avx2.pmul.dq" ||
3307 Name ==
"avx512.pmul.dq.512" ||
3308 Name.starts_with(
"avx512.mask.pmul.dq.")) {
3310 }
else if (Name ==
"sse.cvtsi2ss" || Name ==
"sse2.cvtsi2sd" ||
3311 Name ==
"sse.cvtsi642ss" || Name ==
"sse2.cvtsi642sd") {
3316 }
else if (Name ==
"avx512.cvtusi2sd") {
3321 }
else if (Name ==
"sse2.cvtss2sd") {
3323 Rep = Builder.CreateFPExt(
3326 }
else if (Name ==
"sse2.cvtdq2pd" || Name ==
"sse2.cvtdq2ps" ||
3327 Name ==
"avx.cvtdq2.pd.256" || Name ==
"avx.cvtdq2.ps.256" ||
3328 Name.starts_with(
"avx512.mask.cvtdq2pd.") ||
3329 Name.starts_with(
"avx512.mask.cvtudq2pd.") ||
3330 Name.starts_with(
"avx512.mask.cvtdq2ps.") ||
3331 Name.starts_with(
"avx512.mask.cvtudq2ps.") ||
3332 Name.starts_with(
"avx512.mask.cvtqq2pd.") ||
3333 Name.starts_with(
"avx512.mask.cvtuqq2pd.") ||
3334 Name ==
"avx512.mask.cvtqq2ps.256" ||
3335 Name ==
"avx512.mask.cvtqq2ps.512" ||
3336 Name ==
"avx512.mask.cvtuqq2ps.256" ||
3337 Name ==
"avx512.mask.cvtuqq2ps.512" || Name ==
"sse2.cvtps2pd" ||
3338 Name ==
"avx.cvt.ps2.pd.256" ||
3339 Name ==
"avx512.mask.cvtps2pd.128" ||
3340 Name ==
"avx512.mask.cvtps2pd.256") {
3345 unsigned NumDstElts = DstTy->getNumElements();
3346 if (NumDstElts < SrcTy->getNumElements()) {
3347 assert(NumDstElts == 2 &&
"Unexpected vector size");
3348 Rep = Builder.CreateShuffleVector(Rep, Rep,
ArrayRef<int>{0, 1});
3351 bool IsPS2PD = SrcTy->getElementType()->isFloatTy();
3352 bool IsUnsigned = Name.contains(
"cvtu");
3354 Rep = Builder.CreateFPExt(Rep, DstTy,
"cvtps2pd");
3358 Intrinsic::ID IID = IsUnsigned ? Intrinsic::x86_avx512_uitofp_round
3359 : Intrinsic::x86_avx512_sitofp_round;
3360 Rep = Builder.CreateIntrinsic(IID, {DstTy, SrcTy},
3363 Rep = IsUnsigned ? Builder.CreateUIToFP(Rep, DstTy,
"cvt")
3364 : Builder.CreateSIToFP(Rep, DstTy,
"cvt");
3370 }
else if (Name.starts_with(
"avx512.mask.vcvtph2ps.") ||
3371 Name.starts_with(
"vcvtph2ps.")) {
3375 unsigned NumDstElts = DstTy->getNumElements();
3376 if (NumDstElts != SrcTy->getNumElements()) {
3377 assert(NumDstElts == 4 &&
"Unexpected vector size");
3378 Rep = Builder.CreateShuffleVector(Rep, Rep,
ArrayRef<int>{0, 1, 2, 3});
3380 Rep = Builder.CreateBitCast(
3382 Rep = Builder.CreateFPExt(Rep, DstTy,
"cvtph2ps");
3386 }
else if (Name.starts_with(
"avx512.mask.load")) {
3388 bool Aligned = Name[16] !=
'u';
3391 }
else if (Name.starts_with(
"avx512.mask.expand.load.")) {
3395 ResultTy->getNumElements());
3396 Rep = Builder.CreateIntrinsic(
3397 Intrinsic::masked_expandload, {ResultTy, PtrTy},
3399 }
else if (Name.starts_with(
"avx512.mask.compress.store.")) {
3405 Rep = Builder.CreateIntrinsic(
3406 Intrinsic::masked_compressstore, {ResultTy, PtrTy},
3408 }
else if (Name.starts_with(
"avx512.mask.compress.") ||
3409 Name.starts_with(
"avx512.mask.expand.")) {
3413 ResultTy->getNumElements());
3415 bool IsCompress = Name[12] ==
'c';
3416 Intrinsic::ID IID = IsCompress ? Intrinsic::x86_avx512_mask_compress
3417 : Intrinsic::x86_avx512_mask_expand;
3418 Rep = Builder.CreateIntrinsic(
3420 }
else if (Name.starts_with(
"xop.vpcom")) {
3422 if (Name.ends_with(
"ub") || Name.ends_with(
"uw") || Name.ends_with(
"ud") ||
3423 Name.ends_with(
"uq"))
3425 else if (Name.ends_with(
"b") || Name.ends_with(
"w") ||
3426 Name.ends_with(
"d") || Name.ends_with(
"q"))
3435 Name = Name.substr(9);
3436 if (Name.starts_with(
"lt"))
3438 else if (Name.starts_with(
"le"))
3440 else if (Name.starts_with(
"gt"))
3442 else if (Name.starts_with(
"ge"))
3444 else if (Name.starts_with(
"eq"))
3446 else if (Name.starts_with(
"ne"))
3448 else if (Name.starts_with(
"false"))
3450 else if (Name.starts_with(
"true"))
3457 }
else if (Name.starts_with(
"xop.vpcmov")) {
3459 Value *NotSel = Builder.CreateNot(Sel);
3462 Rep = Builder.CreateOr(Sel0, Sel1);
3463 }
else if (Name.starts_with(
"xop.vprot") || Name.starts_with(
"avx512.prol") ||
3464 Name.starts_with(
"avx512.mask.prol")) {
3466 }
else if (Name.starts_with(
"avx512.pror") ||
3467 Name.starts_with(
"avx512.mask.pror")) {
3469 }
else if (Name.starts_with(
"avx512.vpshld.") ||
3470 Name.starts_with(
"avx512.mask.vpshld") ||
3471 Name.starts_with(
"avx512.maskz.vpshld")) {
3472 bool ZeroMask = Name[11] ==
'z';
3474 }
else if (Name.starts_with(
"avx512.vpshrd.") ||
3475 Name.starts_with(
"avx512.mask.vpshrd") ||
3476 Name.starts_with(
"avx512.maskz.vpshrd")) {
3477 bool ZeroMask = Name[11] ==
'z';
3479 }
else if (Name ==
"sse42.crc32.64.8") {
3482 Rep = Builder.CreateIntrinsic(Intrinsic::x86_sse42_crc32_32_8,
3484 Rep = Builder.CreateZExt(Rep, CI->
getType(),
"");
3485 }
else if (Name.starts_with(
"avx.vbroadcast.s") ||
3486 Name.starts_with(
"avx512.vbroadcast.s")) {
3489 Type *EltTy = VecTy->getElementType();
3490 unsigned EltNum = VecTy->getNumElements();
3494 for (
unsigned I = 0;
I < EltNum; ++
I)
3495 Rep = Builder.CreateInsertElement(Rep,
Load, ConstantInt::get(I32Ty,
I));
3496 }
else if (Name.starts_with(
"sse41.pmovsx") ||
3497 Name.starts_with(
"sse41.pmovzx") ||
3498 Name.starts_with(
"avx2.pmovsx") ||
3499 Name.starts_with(
"avx2.pmovzx") ||
3500 Name.starts_with(
"avx512.mask.pmovsx") ||
3501 Name.starts_with(
"avx512.mask.pmovzx")) {
3503 unsigned NumDstElts = DstTy->getNumElements();
3507 for (
unsigned i = 0; i != NumDstElts; ++i)
3512 bool DoSext = Name.contains(
"pmovsx");
3514 DoSext ? Builder.CreateSExt(SV, DstTy) : Builder.CreateZExt(SV, DstTy);
3519 }
else if (Name ==
"avx512.mask.pmov.qd.256" ||
3520 Name ==
"avx512.mask.pmov.qd.512" ||
3521 Name ==
"avx512.mask.pmov.wb.256" ||
3522 Name ==
"avx512.mask.pmov.wb.512") {
3527 }
else if (Name.starts_with(
"avx.vbroadcastf128") ||
3528 Name ==
"avx2.vbroadcasti128") {
3534 if (NumSrcElts == 2)
3537 Rep = Builder.CreateShuffleVector(
Load,
3539 }
else if (Name.starts_with(
"avx512.mask.shuf.i") ||
3540 Name.starts_with(
"avx512.mask.shuf.f")) {
3545 unsigned ControlBitsMask = NumLanes - 1;
3546 unsigned NumControlBits = NumLanes / 2;
3549 for (
unsigned l = 0; l != NumLanes; ++l) {
3550 unsigned LaneMask = (Imm >> (l * NumControlBits)) & ControlBitsMask;
3552 if (l >= NumLanes / 2)
3553 LaneMask += NumLanes;
3554 for (
unsigned i = 0; i != NumElementsInLane; ++i)
3555 ShuffleMask.push_back(LaneMask * NumElementsInLane + i);
3561 }
else if (Name.starts_with(
"avx512.mask.broadcastf") ||
3562 Name.starts_with(
"avx512.mask.broadcasti")) {
3565 unsigned NumDstElts =
3569 for (
unsigned i = 0; i != NumDstElts; ++i)
3570 ShuffleMask[i] = i % NumSrcElts;
3576 }
else if (Name.starts_with(
"avx2.pbroadcast") ||
3577 Name.starts_with(
"avx2.vbroadcast") ||
3578 Name.starts_with(
"avx512.pbroadcast") ||
3579 Name.starts_with(
"avx512.mask.broadcast.s")) {
3586 Rep = Builder.CreateShuffleVector(
Op, M);
3591 }
else if (Name.starts_with(
"sse2.padds.") ||
3592 Name.starts_with(
"avx2.padds.") ||
3593 Name.starts_with(
"avx512.padds.") ||
3594 Name.starts_with(
"avx512.mask.padds.")) {
3596 }
else if (Name.starts_with(
"sse2.psubs.") ||
3597 Name.starts_with(
"avx2.psubs.") ||
3598 Name.starts_with(
"avx512.psubs.") ||
3599 Name.starts_with(
"avx512.mask.psubs.")) {
3601 }
else if (Name.starts_with(
"sse2.paddus.") ||
3602 Name.starts_with(
"avx2.paddus.") ||
3603 Name.starts_with(
"avx512.mask.paddus.")) {
3605 }
else if (Name.starts_with(
"sse2.psubus.") ||
3606 Name.starts_with(
"avx2.psubus.") ||
3607 Name.starts_with(
"avx512.mask.psubus.")) {
3609 }
else if (Name.starts_with(
"avx512.mask.palignr.")) {
3614 }
else if (Name.starts_with(
"avx512.mask.valign.")) {
3618 }
else if (Name ==
"sse2.psll.dq" || Name ==
"avx2.psll.dq") {
3623 }
else if (Name ==
"sse2.psrl.dq" || Name ==
"avx2.psrl.dq") {
3628 }
else if (Name ==
"sse2.psll.dq.bs" || Name ==
"avx2.psll.dq.bs" ||
3629 Name ==
"avx512.psll.dq.512") {
3633 }
else if (Name ==
"sse2.psrl.dq.bs" || Name ==
"avx2.psrl.dq.bs" ||
3634 Name ==
"avx512.psrl.dq.512") {
3638 }
else if (Name ==
"sse41.pblendw" || Name.starts_with(
"sse41.blendp") ||
3639 Name.starts_with(
"avx.blend.p") || Name ==
"avx2.pblendw" ||
3640 Name.starts_with(
"avx2.pblendd.")) {
3645 unsigned NumElts = VecTy->getNumElements();
3648 for (
unsigned i = 0; i != NumElts; ++i)
3649 Idxs[i] = ((Imm >> (i % 8)) & 1) ? i + NumElts : i;
3651 Rep = Builder.CreateShuffleVector(Op0, Op1, Idxs);
3652 }
else if (Name.starts_with(
"avx.vinsertf128.") ||
3653 Name ==
"avx2.vinserti128" ||
3654 Name.starts_with(
"avx512.mask.insert")) {
3658 unsigned DstNumElts =
3660 unsigned SrcNumElts =
3662 unsigned Scale = DstNumElts / SrcNumElts;
3669 for (
unsigned i = 0; i != SrcNumElts; ++i)
3671 for (
unsigned i = SrcNumElts; i != DstNumElts; ++i)
3672 Idxs[i] = SrcNumElts;
3673 Rep = Builder.CreateShuffleVector(Op1, Idxs);
3687 for (
unsigned i = 0; i != DstNumElts; ++i)
3690 for (
unsigned i = 0; i != SrcNumElts; ++i)
3691 Idxs[i + Imm * SrcNumElts] = i + DstNumElts;
3692 Rep = Builder.CreateShuffleVector(Op0, Rep, Idxs);
3698 }
else if (Name.starts_with(
"avx.vextractf128.") ||
3699 Name ==
"avx2.vextracti128" ||
3700 Name.starts_with(
"avx512.mask.vextract")) {
3703 unsigned DstNumElts =
3705 unsigned SrcNumElts =
3707 unsigned Scale = SrcNumElts / DstNumElts;
3714 for (
unsigned i = 0; i != DstNumElts; ++i) {
3715 Idxs[i] = i + (Imm * DstNumElts);
3717 Rep = Builder.CreateShuffleVector(Op0, Op0, Idxs);
3723 }
else if (Name.starts_with(
"avx512.mask.perm.df.") ||
3724 Name.starts_with(
"avx512.mask.perm.di.")) {
3728 unsigned NumElts = VecTy->getNumElements();
3731 for (
unsigned i = 0; i != NumElts; ++i)
3732 Idxs[i] = (i & ~0x3) + ((Imm >> (2 * (i & 0x3))) & 3);
3734 Rep = Builder.CreateShuffleVector(Op0, Op0, Idxs);
3739 }
else if (Name.starts_with(
"avx.vperm2f128.") || Name ==
"avx2.vperm2i128") {
3751 unsigned HalfSize = NumElts / 2;
3763 unsigned StartIndex = (Imm & 0x01) ? HalfSize : 0;
3764 for (
unsigned i = 0; i < HalfSize; ++i)
3765 ShuffleMask[i] = StartIndex + i;
3768 StartIndex = (Imm & 0x10) ? HalfSize : 0;
3769 for (
unsigned i = 0; i < HalfSize; ++i)
3770 ShuffleMask[i + HalfSize] = NumElts + StartIndex + i;
3772 Rep = Builder.CreateShuffleVector(V0,
V1, ShuffleMask);
3774 }
else if (Name.starts_with(
"avx.vpermil.") || Name ==
"sse2.pshuf.d" ||
3775 Name.starts_with(
"avx512.mask.vpermil.p") ||
3776 Name.starts_with(
"avx512.mask.pshuf.d.")) {
3780 unsigned NumElts = VecTy->getNumElements();
3782 unsigned IdxSize = 64 / VecTy->getScalarSizeInBits();
3783 unsigned IdxMask = ((1 << IdxSize) - 1);
3789 for (
unsigned i = 0; i != NumElts; ++i)
3790 Idxs[i] = ((Imm >> ((i * IdxSize) % 8)) & IdxMask) | (i & ~IdxMask);
3792 Rep = Builder.CreateShuffleVector(Op0, Op0, Idxs);
3797 }
else if (Name ==
"sse2.pshufl.w" ||
3798 Name.starts_with(
"avx512.mask.pshufl.w.")) {
3803 if (Name ==
"sse2.pshufl.w" && NumElts % 8 != 0)
3807 for (
unsigned l = 0; l != NumElts; l += 8) {
3808 for (
unsigned i = 0; i != 4; ++i)
3809 Idxs[i + l] = ((Imm >> (2 * i)) & 0x3) + l;
3810 for (
unsigned i = 4; i != 8; ++i)
3811 Idxs[i + l] = i + l;
3814 Rep = Builder.CreateShuffleVector(Op0, Op0, Idxs);
3819 }
else if (Name ==
"sse2.pshufh.w" ||
3820 Name.starts_with(
"avx512.mask.pshufh.w.")) {
3825 if (Name ==
"sse2.pshufh.w" && NumElts % 8 != 0)
3829 for (
unsigned l = 0; l != NumElts; l += 8) {
3830 for (
unsigned i = 0; i != 4; ++i)
3831 Idxs[i + l] = i + l;
3832 for (
unsigned i = 0; i != 4; ++i)
3833 Idxs[i + l + 4] = ((Imm >> (2 * i)) & 0x3) + 4 + l;
3836 Rep = Builder.CreateShuffleVector(Op0, Op0, Idxs);
3841 }
else if (Name.starts_with(
"avx512.mask.shuf.p")) {
3848 unsigned HalfLaneElts = NumLaneElts / 2;
3851 for (
unsigned i = 0; i != NumElts; ++i) {
3853 Idxs[i] = i - (i % NumLaneElts);
3855 if ((i % NumLaneElts) >= HalfLaneElts)
3859 Idxs[i] += (Imm >> ((i * HalfLaneElts) % 8)) & ((1 << HalfLaneElts) - 1);
3862 Rep = Builder.CreateShuffleVector(Op0, Op1, Idxs);
3866 }
else if (Name.starts_with(
"avx512.mask.movddup") ||
3867 Name.starts_with(
"avx512.mask.movshdup") ||
3868 Name.starts_with(
"avx512.mask.movsldup")) {
3874 if (Name.starts_with(
"avx512.mask.movshdup."))
3878 for (
unsigned l = 0; l != NumElts; l += NumLaneElts)
3879 for (
unsigned i = 0; i != NumLaneElts; i += 2) {
3880 Idxs[i + l + 0] = i + l +
Offset;
3881 Idxs[i + l + 1] = i + l +
Offset;
3884 Rep = Builder.CreateShuffleVector(Op0, Op0, Idxs);
3888 }
else if (Name.starts_with(
"avx512.mask.punpckl") ||
3889 Name.starts_with(
"avx512.mask.unpckl.")) {
3896 for (
int l = 0; l != NumElts; l += NumLaneElts)
3897 for (
int i = 0; i != NumLaneElts; ++i)
3898 Idxs[i + l] = l + (i / 2) + NumElts * (i % 2);
3900 Rep = Builder.CreateShuffleVector(Op0, Op1, Idxs);
3904 }
else if (Name.starts_with(
"avx512.mask.punpckh") ||
3905 Name.starts_with(
"avx512.mask.unpckh.")) {
3912 for (
int l = 0; l != NumElts; l += NumLaneElts)
3913 for (
int i = 0; i != NumLaneElts; ++i)
3914 Idxs[i + l] = (NumLaneElts / 2) + l + (i / 2) + NumElts * (i % 2);
3916 Rep = Builder.CreateShuffleVector(Op0, Op1, Idxs);
3920 }
else if (Name.starts_with(
"avx512.mask.and.") ||
3921 Name.starts_with(
"avx512.mask.pand.")) {
3924 Rep = Builder.CreateAnd(Builder.CreateBitCast(CI->
getArgOperand(0), ITy),
3926 Rep = Builder.CreateBitCast(Rep, FTy);
3929 }
else if (Name.starts_with(
"avx512.mask.andn.") ||
3930 Name.starts_with(
"avx512.mask.pandn.")) {
3933 Rep = Builder.CreateNot(Builder.CreateBitCast(CI->
getArgOperand(0), ITy));
3934 Rep = Builder.CreateAnd(Rep,
3936 Rep = Builder.CreateBitCast(Rep, FTy);
3939 }
else if (Name.starts_with(
"avx512.mask.or.") ||
3940 Name.starts_with(
"avx512.mask.por.")) {
3943 Rep = Builder.CreateOr(Builder.CreateBitCast(CI->
getArgOperand(0), ITy),
3945 Rep = Builder.CreateBitCast(Rep, FTy);
3948 }
else if (Name.starts_with(
"avx512.mask.xor.") ||
3949 Name.starts_with(
"avx512.mask.pxor.")) {
3952 Rep = Builder.CreateXor(Builder.CreateBitCast(CI->
getArgOperand(0), ITy),
3954 Rep = Builder.CreateBitCast(Rep, FTy);
3957 }
else if (Name.starts_with(
"avx512.mask.padd.")) {
3961 }
else if (Name.starts_with(
"avx512.mask.psub.")) {
3965 }
else if (Name.starts_with(
"avx512.mask.pmull.")) {
3969 }
else if (Name.starts_with(
"avx512.mask.add.p")) {
3970 if (Name.ends_with(
".512")) {
3972 if (Name[17] ==
's')
3973 IID = Intrinsic::x86_avx512_add_ps_512;
3975 IID = Intrinsic::x86_avx512_add_pd_512;
3977 Rep = Builder.CreateIntrinsic(
3985 }
else if (Name.starts_with(
"avx512.mask.div.p")) {
3986 if (Name.ends_with(
".512")) {
3988 if (Name[17] ==
's')
3989 IID = Intrinsic::x86_avx512_div_ps_512;
3991 IID = Intrinsic::x86_avx512_div_pd_512;
3993 Rep = Builder.CreateIntrinsic(
4001 }
else if (Name.starts_with(
"avx512.mask.mul.p")) {
4002 if (Name.ends_with(
".512")) {
4004 if (Name[17] ==
's')
4005 IID = Intrinsic::x86_avx512_mul_ps_512;
4007 IID = Intrinsic::x86_avx512_mul_pd_512;
4009 Rep = Builder.CreateIntrinsic(
4017 }
else if (Name.starts_with(
"avx512.mask.sub.p")) {
4018 if (Name.ends_with(
".512")) {
4020 if (Name[17] ==
's')
4021 IID = Intrinsic::x86_avx512_sub_ps_512;
4023 IID = Intrinsic::x86_avx512_sub_pd_512;
4025 Rep = Builder.CreateIntrinsic(
4033 }
else if ((Name.starts_with(
"avx512.mask.max.p") ||
4034 Name.starts_with(
"avx512.mask.min.p")) &&
4035 Name.drop_front(18) ==
".512") {
4036 bool IsDouble = Name[17] ==
'd';
4037 bool IsMin = Name[13] ==
'i';
4039 {Intrinsic::x86_avx512_max_ps_512, Intrinsic::x86_avx512_max_pd_512},
4040 {Intrinsic::x86_avx512_min_ps_512, Intrinsic::x86_avx512_min_pd_512}};
4043 Rep = Builder.CreateIntrinsic(
4048 }
else if (Name.starts_with(
"avx512.mask.lzcnt.")) {
4050 Builder.CreateIntrinsic(Intrinsic::ctlz, CI->
getType(),
4051 {CI->getArgOperand(0), Builder.getInt1(false)});
4054 }
else if (Name.starts_with(
"avx512.mask.psll")) {
4055 bool IsImmediate = Name[16] ==
'i' || (Name.size() > 18 && Name[18] ==
'i');
4056 bool IsVariable = Name[16] ==
'v';
4057 char Size = Name[16] ==
'.' ? Name[17]
4058 : Name[17] ==
'.' ? Name[18]
4059 : Name[18] ==
'.' ? Name[19]
4063 if (IsVariable && Name[17] !=
'.') {
4064 if (
Size ==
'd' && Name[17] ==
'2')
4065 IID = Intrinsic::x86_avx2_psllv_q;
4066 else if (
Size ==
'd' && Name[17] ==
'4')
4067 IID = Intrinsic::x86_avx2_psllv_q_256;
4068 else if (
Size ==
's' && Name[17] ==
'4')
4069 IID = Intrinsic::x86_avx2_psllv_d;
4070 else if (
Size ==
's' && Name[17] ==
'8')
4071 IID = Intrinsic::x86_avx2_psllv_d_256;
4072 else if (
Size ==
'h' && Name[17] ==
'8')
4073 IID = Intrinsic::x86_avx512_psllv_w_128;
4074 else if (
Size ==
'h' && Name[17] ==
'1')
4075 IID = Intrinsic::x86_avx512_psllv_w_256;
4076 else if (Name[17] ==
'3' && Name[18] ==
'2')
4077 IID = Intrinsic::x86_avx512_psllv_w_512;
4080 }
else if (Name.ends_with(
".128")) {
4082 IID = IsImmediate ? Intrinsic::x86_sse2_pslli_d
4083 : Intrinsic::x86_sse2_psll_d;
4084 else if (
Size ==
'q')
4085 IID = IsImmediate ? Intrinsic::x86_sse2_pslli_q
4086 : Intrinsic::x86_sse2_psll_q;
4087 else if (
Size ==
'w')
4088 IID = IsImmediate ? Intrinsic::x86_sse2_pslli_w
4089 : Intrinsic::x86_sse2_psll_w;
4092 }
else if (Name.ends_with(
".256")) {
4094 IID = IsImmediate ? Intrinsic::x86_avx2_pslli_d
4095 : Intrinsic::x86_avx2_psll_d;
4096 else if (
Size ==
'q')
4097 IID = IsImmediate ? Intrinsic::x86_avx2_pslli_q
4098 : Intrinsic::x86_avx2_psll_q;
4099 else if (
Size ==
'w')
4100 IID = IsImmediate ? Intrinsic::x86_avx2_pslli_w
4101 : Intrinsic::x86_avx2_psll_w;
4106 IID = IsImmediate ? Intrinsic::x86_avx512_pslli_d_512
4107 : IsVariable ? Intrinsic::x86_avx512_psllv_d_512
4108 : Intrinsic::x86_avx512_psll_d_512;
4109 else if (
Size ==
'q')
4110 IID = IsImmediate ? Intrinsic::x86_avx512_pslli_q_512
4111 : IsVariable ? Intrinsic::x86_avx512_psllv_q_512
4112 : Intrinsic::x86_avx512_psll_q_512;
4113 else if (
Size ==
'w')
4114 IID = IsImmediate ? Intrinsic::x86_avx512_pslli_w_512
4115 : Intrinsic::x86_avx512_psll_w_512;
4121 }
else if (Name.starts_with(
"avx512.mask.psrl")) {
4122 bool IsImmediate = Name[16] ==
'i' || (Name.size() > 18 && Name[18] ==
'i');
4123 bool IsVariable = Name[16] ==
'v';
4124 char Size = Name[16] ==
'.' ? Name[17]
4125 : Name[17] ==
'.' ? Name[18]
4126 : Name[18] ==
'.' ? Name[19]
4130 if (IsVariable && Name[17] !=
'.') {
4131 if (
Size ==
'd' && Name[17] ==
'2')
4132 IID = Intrinsic::x86_avx2_psrlv_q;
4133 else if (
Size ==
'd' && Name[17] ==
'4')
4134 IID = Intrinsic::x86_avx2_psrlv_q_256;
4135 else if (
Size ==
's' && Name[17] ==
'4')
4136 IID = Intrinsic::x86_avx2_psrlv_d;
4137 else if (
Size ==
's' && Name[17] ==
'8')
4138 IID = Intrinsic::x86_avx2_psrlv_d_256;
4139 else if (
Size ==
'h' && Name[17] ==
'8')
4140 IID = Intrinsic::x86_avx512_psrlv_w_128;
4141 else if (
Size ==
'h' && Name[17] ==
'1')
4142 IID = Intrinsic::x86_avx512_psrlv_w_256;
4143 else if (Name[17] ==
'3' && Name[18] ==
'2')
4144 IID = Intrinsic::x86_avx512_psrlv_w_512;
4147 }
else if (Name.ends_with(
".128")) {
4149 IID = IsImmediate ? Intrinsic::x86_sse2_psrli_d
4150 : Intrinsic::x86_sse2_psrl_d;
4151 else if (
Size ==
'q')
4152 IID = IsImmediate ? Intrinsic::x86_sse2_psrli_q
4153 : Intrinsic::x86_sse2_psrl_q;
4154 else if (
Size ==
'w')
4155 IID = IsImmediate ? Intrinsic::x86_sse2_psrli_w
4156 : Intrinsic::x86_sse2_psrl_w;
4159 }
else if (Name.ends_with(
".256")) {
4161 IID = IsImmediate ? Intrinsic::x86_avx2_psrli_d
4162 : Intrinsic::x86_avx2_psrl_d;
4163 else if (
Size ==
'q')
4164 IID = IsImmediate ? Intrinsic::x86_avx2_psrli_q
4165 : Intrinsic::x86_avx2_psrl_q;
4166 else if (
Size ==
'w')
4167 IID = IsImmediate ? Intrinsic::x86_avx2_psrli_w
4168 : Intrinsic::x86_avx2_psrl_w;
4173 IID = IsImmediate ? Intrinsic::x86_avx512_psrli_d_512
4174 : IsVariable ? Intrinsic::x86_avx512_psrlv_d_512
4175 : Intrinsic::x86_avx512_psrl_d_512;
4176 else if (
Size ==
'q')
4177 IID = IsImmediate ? Intrinsic::x86_avx512_psrli_q_512
4178 : IsVariable ? Intrinsic::x86_avx512_psrlv_q_512
4179 : Intrinsic::x86_avx512_psrl_q_512;
4180 else if (
Size ==
'w')
4181 IID = IsImmediate ? Intrinsic::x86_avx512_psrli_w_512
4182 : Intrinsic::x86_avx512_psrl_w_512;
4188 }
else if (Name.starts_with(
"avx512.mask.psra")) {
4189 bool IsImmediate = Name[16] ==
'i' || (Name.size() > 18 && Name[18] ==
'i');
4190 bool IsVariable = Name[16] ==
'v';
4191 char Size = Name[16] ==
'.' ? Name[17]
4192 : Name[17] ==
'.' ? Name[18]
4193 : Name[18] ==
'.' ? Name[19]
4197 if (IsVariable && Name[17] !=
'.') {
4198 if (
Size ==
's' && Name[17] ==
'4')
4199 IID = Intrinsic::x86_avx2_psrav_d;
4200 else if (
Size ==
's' && Name[17] ==
'8')
4201 IID = Intrinsic::x86_avx2_psrav_d_256;
4202 else if (
Size ==
'h' && Name[17] ==
'8')
4203 IID = Intrinsic::x86_avx512_psrav_w_128;
4204 else if (
Size ==
'h' && Name[17] ==
'1')
4205 IID = Intrinsic::x86_avx512_psrav_w_256;
4206 else if (Name[17] ==
'3' && Name[18] ==
'2')
4207 IID = Intrinsic::x86_avx512_psrav_w_512;
4210 }
else if (Name.ends_with(
".128")) {
4212 IID = IsImmediate ? Intrinsic::x86_sse2_psrai_d
4213 : Intrinsic::x86_sse2_psra_d;
4214 else if (
Size ==
'q')
4215 IID = IsImmediate ? Intrinsic::x86_avx512_psrai_q_128
4216 : IsVariable ? Intrinsic::x86_avx512_psrav_q_128
4217 : Intrinsic::x86_avx512_psra_q_128;
4218 else if (
Size ==
'w')
4219 IID = IsImmediate ? Intrinsic::x86_sse2_psrai_w
4220 : Intrinsic::x86_sse2_psra_w;
4223 }
else if (Name.ends_with(
".256")) {
4225 IID = IsImmediate ? Intrinsic::x86_avx2_psrai_d
4226 : Intrinsic::x86_avx2_psra_d;
4227 else if (
Size ==
'q')
4228 IID = IsImmediate ? Intrinsic::x86_avx512_psrai_q_256
4229 : IsVariable ? Intrinsic::x86_avx512_psrav_q_256
4230 : Intrinsic::x86_avx512_psra_q_256;
4231 else if (
Size ==
'w')
4232 IID = IsImmediate ? Intrinsic::x86_avx2_psrai_w
4233 : Intrinsic::x86_avx2_psra_w;
4238 IID = IsImmediate ? Intrinsic::x86_avx512_psrai_d_512
4239 : IsVariable ? Intrinsic::x86_avx512_psrav_d_512
4240 : Intrinsic::x86_avx512_psra_d_512;
4241 else if (
Size ==
'q')
4242 IID = IsImmediate ? Intrinsic::x86_avx512_psrai_q_512
4243 : IsVariable ? Intrinsic::x86_avx512_psrav_q_512
4244 : Intrinsic::x86_avx512_psra_q_512;
4245 else if (
Size ==
'w')
4246 IID = IsImmediate ? Intrinsic::x86_avx512_psrai_w_512
4247 : Intrinsic::x86_avx512_psra_w_512;
4253 }
else if (Name.starts_with(
"avx512.mask.move.s")) {
4255 }
else if (Name.starts_with(
"avx512.cvtmask2")) {
4257 }
else if (Name.ends_with(
".movntdqa")) {
4261 LoadInst *LI = Builder.CreateAlignedLoad(
4266 }
else if (Name.starts_with(
"fma.vfmadd.") ||
4267 Name.starts_with(
"fma.vfmsub.") ||
4268 Name.starts_with(
"fma.vfnmadd.") ||
4269 Name.starts_with(
"fma.vfnmsub.")) {
4270 bool NegMul = Name[6] ==
'n';
4271 bool NegAcc = NegMul ? Name[8] ==
's' : Name[7] ==
's';
4272 bool IsScalar = NegMul ? Name[12] ==
's' : Name[11] ==
's';
4283 if (NegMul && !IsScalar)
4284 Ops[0] = Builder.CreateFNeg(
Ops[0]);
4285 if (NegMul && IsScalar)
4286 Ops[1] = Builder.CreateFNeg(
Ops[1]);
4288 Ops[2] = Builder.CreateFNeg(
Ops[2]);
4290 Rep = Builder.CreateIntrinsic(Intrinsic::fma,
Ops[0]->
getType(),
Ops);
4294 }
else if (Name.starts_with(
"fma4.vfmadd.s")) {
4302 Rep = Builder.CreateIntrinsic(Intrinsic::fma,
Ops[0]->
getType(),
Ops);
4306 }
else if (Name.starts_with(
"avx512.mask.vfmadd.s") ||
4307 Name.starts_with(
"avx512.maskz.vfmadd.s") ||
4308 Name.starts_with(
"avx512.mask3.vfmadd.s") ||
4309 Name.starts_with(
"avx512.mask3.vfmsub.s") ||
4310 Name.starts_with(
"avx512.mask3.vfnmsub.s")) {
4311 bool IsMask3 = Name[11] ==
'3';
4312 bool IsMaskZ = Name[11] ==
'z';
4314 Name = Name.drop_front(IsMask3 || IsMaskZ ? 13 : 12);
4315 bool NegMul = Name[2] ==
'n';
4316 bool NegAcc = NegMul ? Name[4] ==
's' : Name[3] ==
's';
4322 if (NegMul && (IsMask3 || IsMaskZ))
4323 A = Builder.CreateFNeg(
A);
4324 if (NegMul && !(IsMask3 || IsMaskZ))
4325 B = Builder.CreateFNeg(
B);
4327 C = Builder.CreateFNeg(
C);
4329 A = Builder.CreateExtractElement(
A, (
uint64_t)0);
4330 B = Builder.CreateExtractElement(
B, (
uint64_t)0);
4331 C = Builder.CreateExtractElement(
C, (
uint64_t)0);
4338 if (Name.back() ==
'd')
4339 IID = Intrinsic::x86_avx512_vfmadd_f64;
4341 IID = Intrinsic::x86_avx512_vfmadd_f32;
4342 Rep = Builder.CreateIntrinsic(IID,
Ops);
4344 Rep = Builder.CreateFMA(
A,
B,
C);
4353 if (NegAcc && IsMask3)
4358 Rep = Builder.CreateInsertElement(CI->
getArgOperand(IsMask3 ? 2 : 0), Rep,
4360 }
else if (Name.starts_with(
"avx512.mask.vfmadd.p") ||
4361 Name.starts_with(
"avx512.mask.vfnmadd.p") ||
4362 Name.starts_with(
"avx512.mask.vfnmsub.p") ||
4363 Name.starts_with(
"avx512.mask3.vfmadd.p") ||
4364 Name.starts_with(
"avx512.mask3.vfmsub.p") ||
4365 Name.starts_with(
"avx512.mask3.vfnmsub.p") ||
4366 Name.starts_with(
"avx512.maskz.vfmadd.p")) {
4367 bool IsMask3 = Name[11] ==
'3';
4368 bool IsMaskZ = Name[11] ==
'z';
4370 Name = Name.drop_front(IsMask3 || IsMaskZ ? 13 : 12);
4371 bool NegMul = Name[2] ==
'n';
4372 bool NegAcc = NegMul ? Name[4] ==
's' : Name[3] ==
's';
4378 if (NegMul && (IsMask3 || IsMaskZ))
4379 A = Builder.CreateFNeg(
A);
4380 if (NegMul && !(IsMask3 || IsMaskZ))
4381 B = Builder.CreateFNeg(
B);
4383 C = Builder.CreateFNeg(
C);
4390 if (Name[Name.size() - 5] ==
's')
4391 IID = Intrinsic::x86_avx512_vfmadd_ps_512;
4393 IID = Intrinsic::x86_avx512_vfmadd_pd_512;
4397 Rep = Builder.CreateFMA(
A,
B,
C);
4405 }
else if (Name.starts_with(
"fma.vfmsubadd.p")) {
4409 if (VecWidth == 128 && EltWidth == 32)
4410 IID = Intrinsic::x86_fma_vfmaddsub_ps;
4411 else if (VecWidth == 256 && EltWidth == 32)
4412 IID = Intrinsic::x86_fma_vfmaddsub_ps_256;
4413 else if (VecWidth == 128 && EltWidth == 64)
4414 IID = Intrinsic::x86_fma_vfmaddsub_pd;
4415 else if (VecWidth == 256 && EltWidth == 64)
4416 IID = Intrinsic::x86_fma_vfmaddsub_pd_256;
4422 Ops[2] = Builder.CreateFNeg(
Ops[2]);
4423 Rep = Builder.CreateIntrinsic(IID,
Ops);
4424 }
else if (Name.starts_with(
"avx512.mask.vfmaddsub.p") ||
4425 Name.starts_with(
"avx512.mask3.vfmaddsub.p") ||
4426 Name.starts_with(
"avx512.maskz.vfmaddsub.p") ||
4427 Name.starts_with(
"avx512.mask3.vfmsubadd.p")) {
4428 bool IsMask3 = Name[11] ==
'3';
4429 bool IsMaskZ = Name[11] ==
'z';
4431 Name = Name.drop_front(IsMask3 || IsMaskZ ? 13 : 12);
4432 bool IsSubAdd = Name[3] ==
's';
4436 if (Name[Name.size() - 5] ==
's')
4437 IID = Intrinsic::x86_avx512_vfmaddsub_ps_512;
4439 IID = Intrinsic::x86_avx512_vfmaddsub_pd_512;
4444 Ops[2] = Builder.CreateFNeg(
Ops[2]);
4446 Rep = Builder.CreateIntrinsic(IID,
Ops);
4455 Value *Odd = Builder.CreateCall(FMA,
Ops);
4456 Ops[2] = Builder.CreateFNeg(
Ops[2]);
4457 Value *Even = Builder.CreateCall(FMA,
Ops);
4463 for (
int i = 0; i != NumElts; ++i)
4464 Idxs[i] = i + (i % 2) * NumElts;
4466 Rep = Builder.CreateShuffleVector(Even, Odd, Idxs);
4474 }
else if (Name.starts_with(
"avx512.mask.pternlog.") ||
4475 Name.starts_with(
"avx512.maskz.pternlog.")) {
4476 bool ZeroMask = Name[11] ==
'z';
4480 if (VecWidth == 128 && EltWidth == 32)
4481 IID = Intrinsic::x86_avx512_pternlog_d_128;
4482 else if (VecWidth == 256 && EltWidth == 32)
4483 IID = Intrinsic::x86_avx512_pternlog_d_256;
4484 else if (VecWidth == 512 && EltWidth == 32)
4485 IID = Intrinsic::x86_avx512_pternlog_d_512;
4486 else if (VecWidth == 128 && EltWidth == 64)
4487 IID = Intrinsic::x86_avx512_pternlog_q_128;
4488 else if (VecWidth == 256 && EltWidth == 64)
4489 IID = Intrinsic::x86_avx512_pternlog_q_256;
4490 else if (VecWidth == 512 && EltWidth == 64)
4491 IID = Intrinsic::x86_avx512_pternlog_q_512;
4497 Rep = Builder.CreateIntrinsic(IID, Args);
4501 }
else if (Name.starts_with(
"avx512.mask.vpmadd52") ||
4502 Name.starts_with(
"avx512.maskz.vpmadd52")) {
4503 bool ZeroMask = Name[11] ==
'z';
4504 bool High = Name[20] ==
'h' || Name[21] ==
'h';
4507 if (VecWidth == 128 && !
High)
4508 IID = Intrinsic::x86_avx512_vpmadd52l_uq_128;
4509 else if (VecWidth == 256 && !
High)
4510 IID = Intrinsic::x86_avx512_vpmadd52l_uq_256;
4511 else if (VecWidth == 512 && !
High)
4512 IID = Intrinsic::x86_avx512_vpmadd52l_uq_512;
4513 else if (VecWidth == 128 &&
High)
4514 IID = Intrinsic::x86_avx512_vpmadd52h_uq_128;
4515 else if (VecWidth == 256 &&
High)
4516 IID = Intrinsic::x86_avx512_vpmadd52h_uq_256;
4517 else if (VecWidth == 512 &&
High)
4518 IID = Intrinsic::x86_avx512_vpmadd52h_uq_512;
4524 Rep = Builder.CreateIntrinsic(IID, Args);
4528 }
else if (Name.starts_with(
"avx512.mask.vpermi2var.") ||
4529 Name.starts_with(
"avx512.mask.vpermt2var.") ||
4530 Name.starts_with(
"avx512.maskz.vpermt2var.")) {
4531 bool ZeroMask = Name[11] ==
'z';
4532 bool IndexForm = Name[17] ==
'i';
4534 }
else if (Name.starts_with(
"avx512.mask.vpdpbusd.") ||
4535 Name.starts_with(
"avx512.maskz.vpdpbusd.") ||
4536 Name.starts_with(
"avx512.mask.vpdpbusds.") ||
4537 Name.starts_with(
"avx512.maskz.vpdpbusds.")) {
4538 bool ZeroMask = Name[11] ==
'z';
4539 bool IsSaturating = Name[ZeroMask ? 21 : 20] ==
's';
4542 if (VecWidth == 128 && !IsSaturating)
4543 IID = Intrinsic::x86_avx512_vpdpbusd_128;
4544 else if (VecWidth == 256 && !IsSaturating)
4545 IID = Intrinsic::x86_avx512_vpdpbusd_256;
4546 else if (VecWidth == 512 && !IsSaturating)
4547 IID = Intrinsic::x86_avx512_vpdpbusd_512;
4548 else if (VecWidth == 128 && IsSaturating)
4549 IID = Intrinsic::x86_avx512_vpdpbusds_128;
4550 else if (VecWidth == 256 && IsSaturating)
4551 IID = Intrinsic::x86_avx512_vpdpbusds_256;
4552 else if (VecWidth == 512 && IsSaturating)
4553 IID = Intrinsic::x86_avx512_vpdpbusds_512;
4563 if (Args[1]->
getType()->isVectorTy() &&
4566 ->isIntegerTy(32) &&
4567 Args[2]->
getType()->isVectorTy() &&
4570 ->isIntegerTy(32)) {
4571 Type *NewArgType =
nullptr;
4572 if (VecWidth == 128)
4574 else if (VecWidth == 256)
4576 else if (VecWidth == 512)
4582 Args[1] = Builder.CreateBitCast(Args[1], NewArgType);
4583 Args[2] = Builder.CreateBitCast(Args[2], NewArgType);
4586 Rep = Builder.CreateIntrinsic(IID, Args);
4590 }
else if (Name.starts_with(
"avx512.mask.vpdpwssd.") ||
4591 Name.starts_with(
"avx512.maskz.vpdpwssd.") ||
4592 Name.starts_with(
"avx512.mask.vpdpwssds.") ||
4593 Name.starts_with(
"avx512.maskz.vpdpwssds.")) {
4594 bool ZeroMask = Name[11] ==
'z';
4595 bool IsSaturating = Name[ZeroMask ? 21 : 20] ==
's';
4598 if (VecWidth == 128 && !IsSaturating)
4599 IID = Intrinsic::x86_avx512_vpdpwssd_128;
4600 else if (VecWidth == 256 && !IsSaturating)
4601 IID = Intrinsic::x86_avx512_vpdpwssd_256;
4602 else if (VecWidth == 512 && !IsSaturating)
4603 IID = Intrinsic::x86_avx512_vpdpwssd_512;
4604 else if (VecWidth == 128 && IsSaturating)
4605 IID = Intrinsic::x86_avx512_vpdpwssds_128;
4606 else if (VecWidth == 256 && IsSaturating)
4607 IID = Intrinsic::x86_avx512_vpdpwssds_256;
4608 else if (VecWidth == 512 && IsSaturating)
4609 IID = Intrinsic::x86_avx512_vpdpwssds_512;
4619 if (Args[1]->
getType()->isVectorTy() &&
4622 ->isIntegerTy(32) &&
4623 Args[2]->
getType()->isVectorTy() &&
4626 ->isIntegerTy(32)) {
4627 Type *NewArgType =
nullptr;
4628 if (VecWidth == 128)
4630 else if (VecWidth == 256)
4632 else if (VecWidth == 512)
4638 Args[1] = Builder.CreateBitCast(Args[1], NewArgType);
4639 Args[2] = Builder.CreateBitCast(Args[2], NewArgType);
4642 Rep = Builder.CreateIntrinsic(IID, Args);
4646 }
else if (Name ==
"addcarryx.u32" || Name ==
"addcarryx.u64" ||
4647 Name ==
"addcarry.u32" || Name ==
"addcarry.u64" ||
4648 Name ==
"subborrow.u32" || Name ==
"subborrow.u64") {
4650 if (Name[0] ==
'a' && Name.back() ==
'2')
4651 IID = Intrinsic::x86_addcarry_32;
4652 else if (Name[0] ==
'a' && Name.back() ==
'4')
4653 IID = Intrinsic::x86_addcarry_64;
4654 else if (Name[0] ==
's' && Name.back() ==
'2')
4655 IID = Intrinsic::x86_subborrow_32;
4656 else if (Name[0] ==
's' && Name.back() ==
'4')
4657 IID = Intrinsic::x86_subborrow_64;
4664 Value *NewCall = Builder.CreateIntrinsic(IID, Args);
4667 Value *
Data = Builder.CreateExtractValue(NewCall, 1);
4670 Value *CF = Builder.CreateExtractValue(NewCall, 0);
4674 }
else if (Name.starts_with(
"avx512.mask.") &&
4677 }
else if (Name.starts_with(
"bmi.pdep.")) {
4679 }
else if (Name.starts_with(
"bmi.pext.")) {
4689 if (Name.starts_with(
"neon.bfcvt")) {
4690 if (Name.starts_with(
"neon.bfcvtn2")) {
4692 std::iota(LoMask.
begin(), LoMask.
end(), 0);
4694 std::iota(ConcatMask.
begin(), ConcatMask.
end(), 0);
4695 Value *Inactive = Builder.CreateShuffleVector(CI->
getOperand(0), LoMask);
4698 return Builder.CreateShuffleVector(Inactive, Trunc, ConcatMask);
4699 }
else if (Name.starts_with(
"neon.bfcvtn")) {
4701 std::iota(ConcatMask.
begin(), ConcatMask.
end(), 0);
4705 dbgs() <<
"Trunc: " << *Trunc <<
"\n";
4706 return Builder.CreateShuffleVector(
4709 return Builder.CreateFPTrunc(CI->
getOperand(0),
4712 }
else if (Name.starts_with(
"sve.fcvt")) {
4715 .
Case(
"sve.fcvt.bf16f32", Intrinsic::aarch64_sve_fcvt_bf16f32_v2)
4716 .
Case(
"sve.fcvtnt.bf16f32",
4717 Intrinsic::aarch64_sve_fcvtnt_bf16f32_v2)
4729 if (Args[1]->
getType() != BadPredTy)
4732 Args[1] = Builder.CreateIntrinsic(Intrinsic::aarch64_sve_convert_to_svbool,
4733 BadPredTy, Args[1]);
4734 Args[1] = Builder.CreateIntrinsic(
4735 Intrinsic::aarch64_sve_convert_from_svbool, GoodPredTy, Args[1]);
4737 return Builder.CreateIntrinsic(NewID, Args,
nullptr,
4741 if (Name ==
"neon.vcvtfp2hf")
4742 return Builder.CreateBitCast(
4743 Builder.CreateFPTrunc(
4747 if (Name ==
"neon.vcvthf2fp")
4748 return Builder.CreateFPExt(
4749 Builder.CreateBitCast(
4759 if (Name ==
"mve.vctp64.old") {
4762 Value *VCTP = Builder.CreateIntrinsic(Intrinsic::arm_mve_vctp64, {},
4765 Value *C1 = Builder.CreateIntrinsic(
4766 Intrinsic::arm_mve_pred_v2i,
4768 return Builder.CreateIntrinsic(
4769 Intrinsic::arm_mve_pred_i2v,
4771 }
else if (Name ==
"mve.mull.int.predicated.v2i64.v4i32.v4i1" ||
4772 Name ==
"mve.vqdmull.predicated.v2i64.v4i32.v4i1" ||
4773 Name ==
"mve.vldr.gather.base.predicated.v2i64.v2i64.v4i1" ||
4774 Name ==
"mve.vldr.gather.base.wb.predicated.v2i64.v2i64.v4i1" ||
4776 "mve.vldr.gather.offset.predicated.v2i64.p0i64.v2i64.v4i1" ||
4777 Name ==
"mve.vldr.gather.offset.predicated.v2i64.p0.v2i64.v4i1" ||
4778 Name ==
"mve.vstr.scatter.base.predicated.v2i64.v2i64.v4i1" ||
4779 Name ==
"mve.vstr.scatter.base.wb.predicated.v2i64.v2i64.v4i1" ||
4781 "mve.vstr.scatter.offset.predicated.p0i64.v2i64.v2i64.v4i1" ||
4782 Name ==
"mve.vstr.scatter.offset.predicated.p0.v2i64.v2i64.v4i1" ||
4783 Name ==
"cde.vcx1q.predicated.v2i64.v4i1" ||
4784 Name ==
"cde.vcx1qa.predicated.v2i64.v4i1" ||
4785 Name ==
"cde.vcx2q.predicated.v2i64.v4i1" ||
4786 Name ==
"cde.vcx2qa.predicated.v2i64.v4i1" ||
4787 Name ==
"cde.vcx3q.predicated.v2i64.v4i1" ||
4788 Name ==
"cde.vcx3qa.predicated.v2i64.v4i1") {
4789 std::vector<Type *> Tys;
4793 case Intrinsic::arm_mve_mull_int_predicated:
4794 case Intrinsic::arm_mve_vqdmull_predicated:
4795 case Intrinsic::arm_mve_vldr_gather_base_predicated:
4798 case Intrinsic::arm_mve_vldr_gather_base_wb_predicated:
4799 case Intrinsic::arm_mve_vstr_scatter_base_predicated:
4800 case Intrinsic::arm_mve_vstr_scatter_base_wb_predicated:
4804 case Intrinsic::arm_mve_vldr_gather_offset_predicated:
4808 case Intrinsic::arm_mve_vstr_scatter_offset_predicated:
4812 case Intrinsic::arm_cde_vcx1q_predicated:
4813 case Intrinsic::arm_cde_vcx1qa_predicated:
4814 case Intrinsic::arm_cde_vcx2q_predicated:
4815 case Intrinsic::arm_cde_vcx2qa_predicated:
4816 case Intrinsic::arm_cde_vcx3q_predicated:
4817 case Intrinsic::arm_cde_vcx3qa_predicated:
4824 std::vector<Value *>
Ops;
4826 Type *Ty =
Op->getType();
4827 if (Ty->getScalarSizeInBits() == 1) {
4828 Value *C1 = Builder.CreateIntrinsic(
4829 Intrinsic::arm_mve_pred_v2i,
4831 Op = Builder.CreateIntrinsic(Intrinsic::arm_mve_pred_i2v, {V2I1Ty}, C1);
4836 return Builder.CreateIntrinsic(
ID, Tys,
Ops,
nullptr,
4851 auto UpgradeLegacyWMMAIUIntrinsicCall =
4856 Args.push_back(Builder.getFalse());
4860 F->getParent(),
F->getIntrinsicID(), OverloadTys);
4867 auto *NewCall =
cast<CallInst>(Builder.CreateCall(NewDecl, Args, Bundles));
4872 NewCall->copyMetadata(*CI);
4876 if (
F->getIntrinsicID() == Intrinsic::amdgcn_wmma_i32_16x16x64_iu8) {
4877 assert(CI->
arg_size() == 7 &&
"Legacy int_amdgcn_wmma_i32_16x16x64_iu8 "
4878 "intrinsic should have 7 arguments");
4881 return UpgradeLegacyWMMAIUIntrinsicCall(
F, CI, Builder, {
T1, T2});
4883 if (
F->getIntrinsicID() == Intrinsic::amdgcn_swmmac_i32_16x16x128_iu8) {
4884 assert(CI->
arg_size() == 8 &&
"Legacy int_amdgcn_swmmac_i32_16x16x128_iu8 "
4885 "intrinsic should have 8 arguments");
4890 return UpgradeLegacyWMMAIUIntrinsicCall(
F, CI, Builder, {
T1, T2, T3, T4});
4893 switch (
F->getIntrinsicID()) {
4896 case Intrinsic::amdgcn_wmma_f32_16x16x4_f32:
4897 case Intrinsic::amdgcn_wmma_f32_16x16x32_bf16:
4898 case Intrinsic::amdgcn_wmma_f32_16x16x32_f16:
4899 case Intrinsic::amdgcn_wmma_f16_16x16x32_f16:
4900 case Intrinsic::amdgcn_wmma_bf16_16x16x32_bf16:
4901 case Intrinsic::amdgcn_wmma_bf16f32_16x16x32_bf16: {
4916 if (
F->getIntrinsicID() == Intrinsic::amdgcn_wmma_bf16f32_16x16x32_bf16)
4919 F->getParent(),
F->getIntrinsicID(), Overloads);
4924 auto *NewCall =
cast<CallInst>(Builder.CreateCall(NewDecl, Args, Bundles));
4929 NewCall->copyMetadata(*CI);
4930 NewCall->takeName(CI);
4952 if (NumOperands < 3)
4965 bool IsVolatile =
false;
4969 if (NumOperands > 3)
4974 if (NumOperands > 5) {
4976 IsVolatile = !VolatileArg || !VolatileArg->
isZero();
4990 if (VT->getElementType()->isIntegerTy(16)) {
4993 Val = Builder.CreateBitCast(Val, AsBF16);
5001 Builder.CreateAtomicRMW(RMWOp, Ptr, Val, std::nullopt, Order, SSID);
5003 unsigned AddrSpace = PtrTy->getAddressSpace();
5006 RMW->
setMetadata(
"amdgpu.no.fine.grained.memory", EmptyMD);
5008 RMW->
setMetadata(
"amdgpu.ignore.denormal.mode", EmptyMD);
5013 MDNode *RangeNotPrivate =
5016 RMW->
setMetadata(LLVMContext::MD_noalias_addrspace, RangeNotPrivate);
5022 return Builder.CreateBitCast(RMW, RetTy);
5043 return MAV->getMetadata();
5052 if (Name ==
"label") {
5054 }
else if (Name ==
"assign") {
5061 }
else if (Name ==
"declare") {
5065 }
else if (Name ==
"addr") {
5075 unwrapMAVOp(CI, 1), ExprNode,
nullptr,
nullptr,
nullptr);
5076 }
else if (Name ==
"value") {
5079 unsigned ExprOp = 2;
5094 assert(DR &&
"Unhandled intrinsic kind in upgrade to DbgRecord");
5102 int64_t OffsetVal =
Offset->getSExtValue();
5103 return Builder.CreateIntrinsic(OffsetVal >= 0
5104 ? Intrinsic::vector_splice_left
5105 : Intrinsic::vector_splice_right,
5107 {CI->getArgOperand(0), CI->getArgOperand(1),
5108 Builder.getInt32(std::abs(OffsetVal))});
5113 if (Name.starts_with(
"to.fp16")) {
5115 Builder.CreateFPTrunc(CI->
getArgOperand(0), Builder.getHalfTy());
5116 return Builder.CreateBitCast(Cast, CI->
getType());
5119 if (Name.starts_with(
"from.fp16")) {
5121 Builder.CreateBitCast(CI->
getArgOperand(0), Builder.getHalfTy());
5122 return Builder.CreateFPExt(Cast, CI->
getType());
5133 if (Defaults.empty())
5136 unsigned OldArgCount = CI->
arg_size();
5137 unsigned NewArgCount = NewFn->
arg_size();
5141 if (OldArgCount >= NewArgCount)
5149 if (OldArgCount < FirstDefault)
5154 for (
unsigned Idx = OldArgCount; Idx < NewArgCount; ++Idx) {
5155 assert(Idx >= FirstDefault && Idx - FirstDefault < Defaults.size() &&
5156 "missing argument outside the default range");
5157 Type *ParamTy = NewFT->getParamType(Idx);
5162 NewArgs.
push_back(ConstantInt::get(ParamTy, Defaults[Idx - FirstDefault]));
5168 CallInst *NewCall = Builder.CreateCall(NewFn, NewArgs, OpBundles);
5200 if (!Name.consume_front(
"llvm."))
5203 bool IsX86 = Name.consume_front(
"x86.");
5204 bool IsNVVM = Name.consume_front(
"nvvm.");
5205 bool IsAArch64 = Name.consume_front(
"aarch64.");
5206 bool IsARM = Name.consume_front(
"arm.");
5207 bool IsAMDGCN = Name.consume_front(
"amdgcn.");
5208 bool IsDbg = Name.consume_front(
"dbg.");
5210 (Name.consume_front(
"experimental.vector.splice") ||
5211 Name.consume_front(
"vector.splice")) &&
5212 !(Name.starts_with(
".left") || Name.starts_with(
".right"));
5213 Value *Rep =
nullptr;
5215 if (!IsX86 && Name ==
"stackprotectorcheck") {
5217 }
else if (IsNVVM) {
5221 }
else if (IsAArch64) {
5225 }
else if (IsAMDGCN) {
5229 }
else if (IsOldSplice) {
5231 }
else if (Name.consume_front(
"convert.")) {
5233 }
else if (Name ==
"lifetime.start.i64" || Name ==
"lifetime.end.i64") {
5246 const auto &DefaultCase = [&]() ->
void {
5254 "Unknown function for CallBase upgrade and isn't just a name change");
5262 "Return type must have changed");
5263 assert(OldST->getNumElements() ==
5265 "Must have same number of elements");
5268 CallInst *NewCI = Builder.CreateCall(NewFn, Args);
5271 for (
unsigned Idx = 0; Idx < OldST->getNumElements(); ++Idx) {
5272 Value *Elem = Builder.CreateExtractValue(NewCI, Idx);
5273 Res = Builder.CreateInsertValue(Res, Elem, Idx);
5297 case Intrinsic::arm_neon_vst1:
5298 case Intrinsic::arm_neon_vst2:
5299 case Intrinsic::arm_neon_vst3:
5300 case Intrinsic::arm_neon_vst4:
5301 case Intrinsic::arm_neon_vst2lane:
5302 case Intrinsic::arm_neon_vst3lane:
5303 case Intrinsic::arm_neon_vst4lane: {
5305 NewCall = Builder.CreateCall(NewFn, Args);
5308 case Intrinsic::aarch64_sve_bfmlalb_lane_v2:
5309 case Intrinsic::aarch64_sve_bfmlalt_lane_v2:
5310 case Intrinsic::aarch64_sve_bfdot_lane_v2: {
5315 NewCall = Builder.CreateCall(NewFn, Args);
5318 case Intrinsic::aarch64_sve_ld3_sret:
5319 case Intrinsic::aarch64_sve_ld4_sret:
5320 case Intrinsic::aarch64_sve_ld2_sret: {
5328 Name = Name.substr(5);
5335 unsigned MinElts = RetTy->getMinNumElements() /
N;
5337 Value *NewLdCall = Builder.CreateCall(NewFn, Args);
5339 for (
unsigned I = 0;
I <
N;
I++) {
5340 Value *SRet = Builder.CreateExtractValue(NewLdCall,
I);
5341 Ret = Builder.CreateInsertVector(RetTy, Ret, SRet,
I * MinElts);
5347 case Intrinsic::coro_end_async:
5348 case Intrinsic::coro_end: {
5350 if (NewFn->
getIntrinsicID() == Intrinsic::coro_end && Args.size() == 2)
5352 NewCall = Builder.CreateCall(NewFn, Args);
5357 CI->
getModule(), Intrinsic::coro_is_in_ramp);
5358 Value *InRamp = Builder.CreateCall(IsInRamp);
5368 case Intrinsic::vector_extract: {
5370 Name = Name.substr(5);
5371 if (!Name.starts_with(
"aarch64.sve.tuple.get")) {
5376 unsigned MinElts = RetTy->getMinNumElements();
5379 NewCall = Builder.CreateCall(NewFn, {CI->
getArgOperand(0), NewIdx});
5383 case Intrinsic::vector_insert: {
5385 Name = Name.substr(5);
5386 if (!Name.starts_with(
"aarch64.sve.tuple")) {
5390 if (Name.starts_with(
"aarch64.sve.tuple.set")) {
5395 NewCall = Builder.CreateCall(
5399 if (Name.starts_with(
"aarch64.sve.tuple.create")) {
5405 assert(
N > 1 &&
"Create is expected to be between 2-4");
5408 unsigned MinElts = RetTy->getMinNumElements() /
N;
5409 for (
unsigned I = 0;
I <
N;
I++) {
5411 Ret = Builder.CreateInsertVector(RetTy, Ret, V,
I * MinElts);
5418 case Intrinsic::arm_neon_bfdot:
5419 case Intrinsic::arm_neon_bfmmla:
5420 case Intrinsic::arm_neon_bfmlalb:
5421 case Intrinsic::arm_neon_bfmlalt:
5422 case Intrinsic::aarch64_neon_bfdot:
5423 case Intrinsic::aarch64_neon_bfmmla:
5424 case Intrinsic::aarch64_neon_bfmlalb:
5425 case Intrinsic::aarch64_neon_bfmlalt: {
5428 "Mismatch between function args and call args");
5429 size_t OperandWidth =
5431 assert((OperandWidth == 64 || OperandWidth == 128) &&
5432 "Unexpected operand width");
5434 auto Iter = CI->
args().begin();
5435 Args.push_back(*Iter++);
5436 Args.push_back(Builder.CreateBitCast(*Iter++, NewTy));
5437 Args.push_back(Builder.CreateBitCast(*Iter++, NewTy));
5438 NewCall = Builder.CreateCall(NewFn, Args);
5442 case Intrinsic::bitreverse:
5443 NewCall = Builder.CreateCall(NewFn, {CI->
getArgOperand(0)});
5446 case Intrinsic::ctlz:
5447 case Intrinsic::cttz: {
5454 Builder.CreateCall(NewFn, {CI->
getArgOperand(0), Builder.getFalse()});
5458 case Intrinsic::objectsize: {
5459 Value *NullIsUnknownSize =
5463 NewCall = Builder.CreateCall(
5468 case Intrinsic::ctpop:
5469 NewCall = Builder.CreateCall(NewFn, {CI->
getArgOperand(0)});
5471 case Intrinsic::dbg_value: {
5473 Name = Name.substr(5);
5475 if (Name.starts_with(
"dbg.addr")) {
5489 if (
Offset->isNullValue()) {
5490 NewCall = Builder.CreateCall(
5499 case Intrinsic::ptr_annotation:
5507 NewCall = Builder.CreateCall(
5516 case Intrinsic::var_annotation:
5523 NewCall = Builder.CreateCall(
5532 case Intrinsic::riscv_aes32dsi:
5533 case Intrinsic::riscv_aes32dsmi:
5534 case Intrinsic::riscv_aes32esi:
5535 case Intrinsic::riscv_aes32esmi:
5536 case Intrinsic::riscv_sm4ks:
5537 case Intrinsic::riscv_sm4ed: {
5547 Arg0 = Builder.CreateTrunc(Arg0, Builder.getInt32Ty());
5548 Arg1 = Builder.CreateTrunc(Arg1, Builder.getInt32Ty());
5554 NewCall = Builder.CreateCall(NewFn, {Arg0, Arg1, Arg2});
5555 Value *Res = NewCall;
5557 Res = Builder.CreateIntCast(NewCall, CI->
getType(),
true);
5563 case Intrinsic::nvvm_mapa_shared_cluster: {
5567 Value *Res = NewCall;
5568 Res = Builder.CreateAddrSpaceCast(
5575 case Intrinsic::nvvm_cp_async_bulk_global_to_shared_cluster:
5576 case Intrinsic::nvvm_cp_async_bulk_shared_cta_to_cluster: {
5579 Args[0] = Builder.CreateAddrSpaceCast(
5582 NewCall = Builder.CreateCall(NewFn, Args);
5588 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_im2col_3d:
5589 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_im2col_4d:
5590 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_im2col_5d:
5591 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_1d:
5592 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_2d:
5593 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_3d:
5594 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_4d:
5595 case Intrinsic::nvvm_cp_async_bulk_tensor_g2s_tile_5d: {
5602 Args[0] = Builder.CreateAddrSpaceCast(
5611 Args.push_back(ConstantInt::get(Builder.getInt32Ty(), 0));
5613 NewCall = Builder.CreateCall(NewFn, Args);
5619 case Intrinsic::riscv_sha256sig0:
5620 case Intrinsic::riscv_sha256sig1:
5621 case Intrinsic::riscv_sha256sum0:
5622 case Intrinsic::riscv_sha256sum1:
5623 case Intrinsic::riscv_sm3p0:
5624 case Intrinsic::riscv_sm3p1: {
5631 Builder.CreateTrunc(CI->
getArgOperand(0), Builder.getInt32Ty());
5633 NewCall = Builder.CreateCall(NewFn, Arg);
5635 Builder.CreateIntCast(NewCall, CI->
getType(),
true);
5642 case Intrinsic::x86_xop_vfrcz_ss:
5643 case Intrinsic::x86_xop_vfrcz_sd:
5644 NewCall = Builder.CreateCall(NewFn, {CI->
getArgOperand(1)});
5647 case Intrinsic::x86_xop_vpermil2pd:
5648 case Intrinsic::x86_xop_vpermil2ps:
5649 case Intrinsic::x86_xop_vpermil2pd_256:
5650 case Intrinsic::x86_xop_vpermil2ps_256: {
5654 Args[2] = Builder.CreateBitCast(Args[2], IntIdxTy);
5655 NewCall = Builder.CreateCall(NewFn, Args);
5659 case Intrinsic::x86_sse41_ptestc:
5660 case Intrinsic::x86_sse41_ptestz:
5661 case Intrinsic::x86_sse41_ptestnzc: {
5675 Value *BC0 = Builder.CreateBitCast(Arg0, NewVecTy,
"cast");
5676 Value *BC1 = Builder.CreateBitCast(Arg1, NewVecTy,
"cast");
5678 NewCall = Builder.CreateCall(NewFn, {BC0, BC1});
5682 case Intrinsic::x86_rdtscp: {
5688 NewCall = Builder.CreateCall(NewFn);
5690 Value *
Data = Builder.CreateExtractValue(NewCall, 1);
5693 Value *TSC = Builder.CreateExtractValue(NewCall, 0);
5701 case Intrinsic::x86_sse41_insertps:
5702 case Intrinsic::x86_sse41_dppd:
5703 case Intrinsic::x86_sse41_dpps:
5704 case Intrinsic::x86_sse41_mpsadbw:
5705 case Intrinsic::x86_avx_dp_ps_256:
5706 case Intrinsic::x86_avx2_mpsadbw: {
5712 Args.back() = Builder.CreateTrunc(Args.back(),
Type::getInt8Ty(
C),
"trunc");
5713 NewCall = Builder.CreateCall(NewFn, Args);
5717 case Intrinsic::x86_avx512_mask_cmp_pd_128:
5718 case Intrinsic::x86_avx512_mask_cmp_pd_256:
5719 case Intrinsic::x86_avx512_mask_cmp_pd_512:
5720 case Intrinsic::x86_avx512_mask_cmp_ps_128:
5721 case Intrinsic::x86_avx512_mask_cmp_ps_256:
5722 case Intrinsic::x86_avx512_mask_cmp_ps_512: {
5728 NewCall = Builder.CreateCall(NewFn, Args);
5737 case Intrinsic::x86_avx512bf16_cvtne2ps2bf16_128:
5738 case Intrinsic::x86_avx512bf16_cvtne2ps2bf16_256:
5739 case Intrinsic::x86_avx512bf16_cvtne2ps2bf16_512:
5740 case Intrinsic::x86_avx512bf16_mask_cvtneps2bf16_128:
5741 case Intrinsic::x86_avx512bf16_cvtneps2bf16_256:
5742 case Intrinsic::x86_avx512bf16_cvtneps2bf16_512: {
5746 Intrinsic::x86_avx512bf16_mask_cvtneps2bf16_128)
5747 Args[1] = Builder.CreateBitCast(
5750 NewCall = Builder.CreateCall(NewFn, Args);
5751 Value *Res = Builder.CreateBitCast(
5759 case Intrinsic::x86_avx512bf16_dpbf16ps_128:
5760 case Intrinsic::x86_avx512bf16_dpbf16ps_256:
5761 case Intrinsic::x86_avx512bf16_dpbf16ps_512:{
5765 Args[1] = Builder.CreateBitCast(
5767 Args[2] = Builder.CreateBitCast(
5770 NewCall = Builder.CreateCall(NewFn, Args);
5774 case Intrinsic::thread_pointer: {
5775 NewCall = Builder.CreateCall(NewFn, {});
5779 case Intrinsic::memcpy:
5780 case Intrinsic::memmove:
5781 case Intrinsic::memset: {
5797 NewCall = Builder.CreateCall(NewFn, Args);
5799 AttributeList NewAttrs = AttributeList::get(
5800 C, OldAttrs.getFnAttrs(), OldAttrs.getRetAttrs(),
5801 {OldAttrs.getParamAttrs(0), OldAttrs.getParamAttrs(1),
5802 OldAttrs.getParamAttrs(2), OldAttrs.getParamAttrs(4)});
5807 MemCI->setDestAlignment(
Align->getMaybeAlignValue());
5810 MTI->setSourceAlignment(
Align->getMaybeAlignValue());
5814 case Intrinsic::masked_load:
5815 case Intrinsic::masked_gather:
5816 case Intrinsic::masked_store:
5817 case Intrinsic::masked_scatter: {
5823 auto GetMaybeAlign = [](
Value *
Op) {
5833 auto GetAlign = [&](
Value *
Op) {
5842 case Intrinsic::masked_load:
5843 NewCall = Builder.CreateMaskedLoad(
5847 case Intrinsic::masked_gather:
5848 NewCall = Builder.CreateMaskedGather(
5854 case Intrinsic::masked_store:
5855 NewCall = Builder.CreateMaskedStore(
5859 case Intrinsic::masked_scatter:
5860 NewCall = Builder.CreateMaskedScatter(
5862 DL.getValueOrABITypeAlignment(
5876 case Intrinsic::lifetime_start:
5877 case Intrinsic::lifetime_end: {
5889 NewCall = Builder.CreateLifetimeStart(Ptr);
5891 NewCall = Builder.CreateLifetimeEnd(Ptr);
5900 case Intrinsic::x86_avx512_vpdpbusd_128:
5901 case Intrinsic::x86_avx512_vpdpbusd_256:
5902 case Intrinsic::x86_avx512_vpdpbusd_512:
5903 case Intrinsic::x86_avx512_vpdpbusds_128:
5904 case Intrinsic::x86_avx512_vpdpbusds_256:
5905 case Intrinsic::x86_avx512_vpdpbusds_512:
5906 case Intrinsic::x86_avx2_vpdpbssd_128:
5907 case Intrinsic::x86_avx2_vpdpbssd_256:
5908 case Intrinsic::x86_avx10_vpdpbssd_512:
5909 case Intrinsic::x86_avx2_vpdpbssds_128:
5910 case Intrinsic::x86_avx2_vpdpbssds_256:
5911 case Intrinsic::x86_avx10_vpdpbssds_512:
5912 case Intrinsic::x86_avx2_vpdpbsud_128:
5913 case Intrinsic::x86_avx2_vpdpbsud_256:
5914 case Intrinsic::x86_avx10_vpdpbsud_512:
5915 case Intrinsic::x86_avx2_vpdpbsuds_128:
5916 case Intrinsic::x86_avx2_vpdpbsuds_256:
5917 case Intrinsic::x86_avx10_vpdpbsuds_512:
5918 case Intrinsic::x86_avx2_vpdpbuud_128:
5919 case Intrinsic::x86_avx2_vpdpbuud_256:
5920 case Intrinsic::x86_avx10_vpdpbuud_512:
5921 case Intrinsic::x86_avx2_vpdpbuuds_128:
5922 case Intrinsic::x86_avx2_vpdpbuuds_256:
5923 case Intrinsic::x86_avx10_vpdpbuuds_512: {
5928 Args[1] = Builder.CreateBitCast(Args[1], NewArgType);
5929 Args[2] = Builder.CreateBitCast(Args[2], NewArgType);
5931 NewCall = Builder.CreateCall(NewFn, Args);
5934 case Intrinsic::x86_avx512_vpdpwssd_128:
5935 case Intrinsic::x86_avx512_vpdpwssd_256:
5936 case Intrinsic::x86_avx512_vpdpwssd_512:
5937 case Intrinsic::x86_avx512_vpdpwssds_128:
5938 case Intrinsic::x86_avx512_vpdpwssds_256:
5939 case Intrinsic::x86_avx512_vpdpwssds_512:
5940 case Intrinsic::x86_avx2_vpdpwsud_128:
5941 case Intrinsic::x86_avx2_vpdpwsud_256:
5942 case Intrinsic::x86_avx10_vpdpwsud_512:
5943 case Intrinsic::x86_avx2_vpdpwsuds_128:
5944 case Intrinsic::x86_avx2_vpdpwsuds_256:
5945 case Intrinsic::x86_avx10_vpdpwsuds_512:
5946 case Intrinsic::x86_avx2_vpdpwusd_128:
5947 case Intrinsic::x86_avx2_vpdpwusd_256:
5948 case Intrinsic::x86_avx10_vpdpwusd_512:
5949 case Intrinsic::x86_avx2_vpdpwusds_128:
5950 case Intrinsic::x86_avx2_vpdpwusds_256:
5951 case Intrinsic::x86_avx10_vpdpwusds_512:
5952 case Intrinsic::x86_avx2_vpdpwuud_128:
5953 case Intrinsic::x86_avx2_vpdpwuud_256:
5954 case Intrinsic::x86_avx10_vpdpwuud_512:
5955 case Intrinsic::x86_avx2_vpdpwuuds_128:
5956 case Intrinsic::x86_avx2_vpdpwuuds_256:
5957 case Intrinsic::x86_avx10_vpdpwuuds_512:
5962 Args[1] = Builder.CreateBitCast(Args[1], NewArgType);
5963 Args[2] = Builder.CreateBitCast(Args[2], NewArgType);
5965 NewCall = Builder.CreateCall(NewFn, Args);
5968 assert(NewCall &&
"Should have either set this variable or returned through "
5969 "the default case");
5976 assert(
F &&
"Illegal attempt to upgrade a non-existent intrinsic.");
5990 F->eraseFromParent();
5996 if (NumOperands == 0)
6004 if (NumOperands == 3) {
6008 Metadata *Elts2[] = {ScalarType, ScalarType,
6022 if (
Opc != Instruction::BitCast)
6026 Type *SrcTy = V->getType();
6043 if (
Opc != Instruction::BitCast)
6046 Type *SrcTy =
C->getType();
6063 if (Flag.getNumOperands() < 3)
6064 return std::nullopt;
6066 return Name->getString();
6067 return std::nullopt;
6081 if (
NamedMDNode *ModFlags = M.getModuleFlagsMetadata()) {
6082 auto OpIt =
find_if(ModFlags->operands(), [](
const MDNode *Flag) {
6083 if (auto Name = getModuleFlagNameSafely(*Flag))
6084 return *Name ==
"Debug Info Version";
6087 if (OpIt != ModFlags->op_end()) {
6088 const MDOperand &ValOp = (*OpIt)->getOperand(2);
6095 bool BrokenDebugInfo =
false;
6098 if (!BrokenDebugInfo)
6104 M.getContext().diagnose(Diag);
6111 M.getContext().diagnose(DiagVersion);
6121 StringRef Vect3[3] = {DefaultValue, DefaultValue, DefaultValue};
6124 if (
F->hasFnAttribute(Attr)) {
6127 StringRef S =
F->getFnAttribute(Attr).getValueAsString();
6129 auto [Part, Rest] = S.
split(
',');
6135 const unsigned Dim = DimC -
'x';
6136 assert(Dim < 3 &&
"Unexpected dim char");
6146 F->addFnAttr(Attr, NewAttr);
6150 return S ==
"x" || S ==
"y" || S ==
"z";
6155 if (K ==
"kernel") {
6167 const unsigned Idx = (AlignIdxValuePair >> 16);
6168 const Align StackAlign =
Align(AlignIdxValuePair & 0xFFFF);
6173 if (K ==
"maxclusterrank" || K ==
"cluster_max_blocks") {
6178 if (K ==
"minctasm") {
6183 if (K ==
"maxnreg") {
6188 if (K.consume_front(
"maxntid") &&
isXYZ(K)) {
6192 if (K.consume_front(
"reqntid") &&
isXYZ(K)) {
6196 if (K.consume_front(
"cluster_dim_") &&
isXYZ(K)) {
6200 if (K ==
"grid_constant") {
6215 NamedMDNode *NamedMD = M.getNamedMetadata(
"nvvm.annotations");
6222 if (!SeenNodes.
insert(MD).second)
6229 assert((MD->getNumOperands() % 2) == 1 &&
"Invalid number of operands");
6236 for (
unsigned j = 1, je = MD->getNumOperands(); j < je; j += 2) {
6238 const MDOperand &V = MD->getOperand(j + 1);
6241 NewOperands.
append({K, V});
6244 if (NewOperands.
size() > 1)
6257 const char *MarkerKey =
"clang.arc.retainAutoreleasedReturnValueMarker";
6258 NamedMDNode *ModRetainReleaseMarker = M.getNamedMetadata(MarkerKey);
6259 if (ModRetainReleaseMarker) {
6265 ID->getString().split(ValueComp,
"#");
6266 if (ValueComp.
size() == 2) {
6267 std::string NewValue = ValueComp[0].str() +
";" + ValueComp[1].str();
6271 M.eraseNamedMetadata(ModRetainReleaseMarker);
6282 auto UpgradeToIntrinsic = [&](
const char *OldFunc,
6308 bool InvalidCast =
false;
6310 for (
unsigned I = 0, E = CI->
arg_size();
I != E; ++
I) {
6323 Arg = Builder.CreateBitCast(Arg, NewFuncTy->
getParamType(
I));
6325 Args.push_back(Arg);
6332 CallInst *NewCall = Builder.CreateCall(NewFuncTy, NewFn, Args);
6337 Value *NewRetVal = Builder.CreateBitCast(NewCall, CI->
getType());
6350 UpgradeToIntrinsic(
"clang.arc.use", llvm::Intrinsic::objc_clang_arc_use);
6358 std::pair<const char *, llvm::Intrinsic::ID> RuntimeFuncs[] = {
6359 {
"objc_autorelease", llvm::Intrinsic::objc_autorelease},
6360 {
"objc_autoreleasePoolPop", llvm::Intrinsic::objc_autoreleasePoolPop},
6361 {
"objc_autoreleasePoolPush", llvm::Intrinsic::objc_autoreleasePoolPush},
6362 {
"objc_autoreleaseReturnValue",
6363 llvm::Intrinsic::objc_autoreleaseReturnValue},
6364 {
"objc_copyWeak", llvm::Intrinsic::objc_copyWeak},
6365 {
"objc_destroyWeak", llvm::Intrinsic::objc_destroyWeak},
6366 {
"objc_initWeak", llvm::Intrinsic::objc_initWeak},
6367 {
"objc_loadWeak", llvm::Intrinsic::objc_loadWeak},
6368 {
"objc_loadWeakRetained", llvm::Intrinsic::objc_loadWeakRetained},
6369 {
"objc_moveWeak", llvm::Intrinsic::objc_moveWeak},
6370 {
"objc_release", llvm::Intrinsic::objc_release},
6371 {
"objc_retain", llvm::Intrinsic::objc_retain},
6372 {
"objc_retainAutorelease", llvm::Intrinsic::objc_retainAutorelease},
6373 {
"objc_retainAutoreleaseReturnValue",
6374 llvm::Intrinsic::objc_retainAutoreleaseReturnValue},
6375 {
"objc_retainAutoreleasedReturnValue",
6376 llvm::Intrinsic::objc_retainAutoreleasedReturnValue},
6377 {
"objc_retainBlock", llvm::Intrinsic::objc_retainBlock},
6378 {
"objc_storeStrong", llvm::Intrinsic::objc_storeStrong},
6379 {
"objc_storeWeak", llvm::Intrinsic::objc_storeWeak},
6380 {
"objc_unsafeClaimAutoreleasedReturnValue",
6381 llvm::Intrinsic::objc_unsafeClaimAutoreleasedReturnValue},
6382 {
"objc_retainedObject", llvm::Intrinsic::objc_retainedObject},
6383 {
"objc_unretainedObject", llvm::Intrinsic::objc_unretainedObject},
6384 {
"objc_unretainedPointer", llvm::Intrinsic::objc_unretainedPointer},
6385 {
"objc_retain_autorelease", llvm::Intrinsic::objc_retain_autorelease},
6386 {
"objc_sync_enter", llvm::Intrinsic::objc_sync_enter},
6387 {
"objc_sync_exit", llvm::Intrinsic::objc_sync_exit},
6388 {
"objc_arc_annotation_topdown_bbstart",
6389 llvm::Intrinsic::objc_arc_annotation_topdown_bbstart},
6390 {
"objc_arc_annotation_topdown_bbend",
6391 llvm::Intrinsic::objc_arc_annotation_topdown_bbend},
6392 {
"objc_arc_annotation_bottomup_bbstart",
6393 llvm::Intrinsic::objc_arc_annotation_bottomup_bbstart},
6394 {
"objc_arc_annotation_bottomup_bbend",
6395 llvm::Intrinsic::objc_arc_annotation_bottomup_bbend}};
6397 for (
auto &
I : RuntimeFuncs)
6398 UpgradeToIntrinsic(
I.first,
I.second);
6422 std::optional<bool> UseAddressDisc;
6425 if (
const NamedMDNode *ModFlags = M.getModuleFlagsMetadata()) {
6426 for (
const MDNode *Flag : ModFlags->operands()) {
6428 if (Name && (*Name ==
"ptrauth-init-fini" ||
6429 *Name ==
"ptrauth-init-fini-address-discrimination"))
6434 auto UpgradeSinglePointer = [&UseAddressDisc](
Constant *CV) ->
Constant * {
6435 constexpr unsigned ExpectedConstDisc = 0xD9D4;
6436 constexpr unsigned ExpectedAddressMarker = 1;
6439 if (!CPA || !CPA->getDiscriminator()->equalsInt(ExpectedConstDisc))
6442 bool HasAddressDisc;
6443 if (!CPA->hasAddressDiscriminator())
6444 HasAddressDisc =
false;
6445 else if (CPA->hasSpecialAddressDiscriminator(ExpectedAddressMarker))
6446 HasAddressDisc =
true;
6450 if (UseAddressDisc && *UseAddressDisc != HasAddressDisc)
6453 UseAddressDisc = HasAddressDisc;
6454 return CPA->getPointer();
6458 using PendingUpgrade = std::pair<GlobalVariable *, Constant *>;
6461 for (
const char *Name : {
"llvm.global_ctors",
"llvm.global_dtors"}) {
6463 if (!GV || !GV->hasInitializer())
6467 if (!OldStructorsArray || OldStructorsArray->getNumOperands() == 0)
6470 std::vector<Constant *> NewStructors;
6471 NewStructors.reserve(OldStructorsArray->getNumOperands());
6473 for (
Use &U : OldStructorsArray->operands()) {
6482 Func = UpgradeSinglePointer(Func);
6486 NewStructors.push_back(
6495 if (GlobalArraysToUpgrade.
empty())
6497 assert(UseAddressDisc.has_value());
6499 for (
auto [GV, NewInit] : GlobalArraysToUpgrade)
6500 GV->setInitializer(NewInit);
6503 M.addModuleFlag(
Module::Error,
"ptrauth-init-fini-address-discrimination",
6513 NamedMDNode *ModFlags = M.getModuleFlagsMetadata();
6517 bool HasObjCFlag =
false, HasClassProperties =
false;
6518 bool HasSwiftVersionFlag =
false;
6519 uint8_t SwiftMajorVersion, SwiftMinorVersion;
6526 if (
Op->getNumOperands() != 3)
6540 if (
ID->getString() ==
"Objective-C Image Info Version")
6542 if (
ID->getString() ==
"Objective-C Class Properties")
6543 HasClassProperties =
true;
6545 if (
ID->getString() ==
"PIC Level") {
6546 if (
auto *Behavior =
6548 uint64_t V = Behavior->getLimitedValue();
6554 if (
ID->getString() ==
"PIE Level")
6555 if (
auto *Behavior =
6562 if (
ID->getString() ==
"branch-target-enforcement" ||
6563 ID->getString().starts_with(
"sign-return-address")) {
6564 if (
auto *Behavior =
6570 Op->getOperand(1),
Op->getOperand(2)};
6580 if (
ID->getString() ==
"Objective-C Image Info Section") {
6583 Value->getString().split(ValueComp,
" ");
6584 if (ValueComp.
size() != 1) {
6585 std::string NewValue;
6586 for (
auto &S : ValueComp)
6587 NewValue += S.str();
6598 if (
ID->getString() ==
"Objective-C Garbage Collection") {
6601 assert(Md->getValue() &&
"Expected non-empty metadata");
6602 auto Type = Md->getValue()->getType();
6605 unsigned Val = Md->getValue()->getUniqueInteger().getZExtValue();
6606 if ((Val & 0xff) != Val) {
6607 HasSwiftVersionFlag =
true;
6608 SwiftABIVersion = (Val & 0xff00) >> 8;
6609 SwiftMajorVersion = (Val & 0xff000000) >> 24;
6610 SwiftMinorVersion = (Val & 0xff0000) >> 16;
6621 if (
ID->getString() ==
"amdgpu_code_object_version") {
6624 MDString::get(M.getContext(),
"amdhsa_code_object_version"),
6636 if (HasObjCFlag && !HasClassProperties) {
6642 if (HasSwiftVersionFlag) {
6646 ConstantInt::get(Int8Ty, SwiftMajorVersion));
6648 ConstantInt::get(Int8Ty, SwiftMinorVersion));
6656 NamedMDNode *CFIConsts = M.getNamedMetadata(
"cfi.functions");
6660 auto MatchesVersion = [](
const MDNode *
Op) {
6661 return Op->getNumOperands() >= 3 &&
6675 assert(!MatchesVersion(
Op) &&
"Unexpected mix of CFIConstant formats");
6676 assert(
Op->getNumOperands() >= 2 &&
6677 "Expected at least 2 operands - name and linkage type");
6689 for (
unsigned J = 2, EJ =
Op->getNumOperands(); J != EJ; ++J)
6700 auto TrimSpaces = [](
StringRef Section) -> std::string {
6702 Section.split(Components,
',');
6707 for (
auto Component : Components)
6708 OS <<
',' << Component.trim();
6713 for (
auto &GV : M.globals()) {
6714 if (!GV.hasSection())
6719 if (!Section.starts_with(
"__DATA, __objc_catlist"))
6724 GV.setSection(TrimSpaces(Section));
6740struct StrictFPUpgradeVisitor :
public InstVisitor<StrictFPUpgradeVisitor> {
6741 StrictFPUpgradeVisitor() =
default;
6744 if (!
Call.isStrictFP())
6750 Call.removeFnAttr(Attribute::StrictFP);
6751 Call.addFnAttr(Attribute::NoBuiltin);
6756struct AMDGPUUnsafeFPAtomicsUpgradeVisitor
6757 :
public InstVisitor<AMDGPUUnsafeFPAtomicsUpgradeVisitor> {
6758 AMDGPUUnsafeFPAtomicsUpgradeVisitor() =
default;
6760 void visitAtomicRMWInst(AtomicRMWInst &RMW) {
6775 if (!
F.isDeclaration() && !
F.hasFnAttribute(Attribute::StrictFP)) {
6776 StrictFPUpgradeVisitor SFPV;
6781 F.removeRetAttrs(AttributeFuncs::typeIncompatible(
6782 F.getReturnType(),
F.getAttributes().getRetAttrs()));
6783 for (
auto &Arg :
F.args())
6785 AttributeFuncs::typeIncompatible(Arg.getType(), Arg.getAttributes()));
6787 bool AddingAttrs =
false, RemovingAttrs =
false;
6788 AttrBuilder AttrsToAdd(
F.getContext());
6793 if (
Attribute A =
F.getFnAttribute(
"implicit-section-name");
6794 A.isValid() &&
A.isStringAttribute()) {
6795 F.setSection(
A.getValueAsString());
6797 RemovingAttrs =
true;
6801 A.isValid() &&
A.isStringAttribute()) {
6804 AddingAttrs = RemovingAttrs =
true;
6807 if (
Attribute A =
F.getFnAttribute(
"uniform-work-group-size");
6808 A.isValid() &&
A.isStringAttribute() && !
A.getValueAsString().empty()) {
6810 RemovingAttrs =
true;
6811 if (
A.getValueAsString() ==
"true") {
6812 AttrsToAdd.addAttribute(
"uniform-work-group-size");
6821 if (
Attribute A =
F.getFnAttribute(
"amdgpu-unsafe-fp-atomics");
6824 if (
A.getValueAsBool()) {
6825 AMDGPUUnsafeFPAtomicsUpgradeVisitor Visitor;
6831 AttrsToRemove.
addAttribute(
"amdgpu-unsafe-fp-atomics");
6832 RemovingAttrs =
true;
6839 bool HandleDenormalMode =
false;
6841 if (
Attribute Attr =
F.getFnAttribute(
"denormal-fp-math"); Attr.isValid()) {
6844 DenormalFPMath = ParsedMode;
6846 AddingAttrs = RemovingAttrs =
true;
6847 HandleDenormalMode =
true;
6851 if (
Attribute Attr =
F.getFnAttribute(
"denormal-fp-math-f32");
6855 DenormalFPMathF32 = ParsedMode;
6857 AddingAttrs = RemovingAttrs =
true;
6858 HandleDenormalMode =
true;
6862 if (HandleDenormalMode)
6863 AttrsToAdd.addDenormalFPEnvAttr(
6867 F.removeFnAttrs(AttrsToRemove);
6870 F.addFnAttrs(AttrsToAdd);
6876 if (!
F.hasFnAttribute(FnAttrName))
6877 F.addFnAttr(FnAttrName,
Value);
6884 if (!
F.hasFnAttribute(FnAttrName)) {
6886 F.addFnAttr(FnAttrName);
6888 auto A =
F.getFnAttribute(FnAttrName);
6889 if (
"false" ==
A.getValueAsString())
6890 F.removeFnAttr(FnAttrName);
6891 else if (
"true" ==
A.getValueAsString()) {
6892 F.removeFnAttr(FnAttrName);
6893 F.addFnAttr(FnAttrName);
6899 Triple T(M.getTargetTriple());
6900 if (!
T.isThumb() && !
T.isARM() && !
T.isAArch64())
6910 NamedMDNode *ModFlags = M.getModuleFlagsMetadata();
6914 if (
Op->getNumOperands() != 3)
6923 uint64_t *ValPtr = IDStr ==
"branch-target-enforcement" ? &BTEValue
6924 : IDStr ==
"branch-protection-pauth-lr" ? &BPPLRValue
6925 : IDStr ==
"guarded-control-stack" ? &GCSValue
6926 : IDStr ==
"sign-return-address" ? &SRAValue
6927 : IDStr ==
"sign-return-address-all" ? &SRAALLValue
6928 : IDStr ==
"sign-return-address-with-bkey"
6934 *ValPtr = CI->getZExtValue();
6940 bool BTE = BTEValue == 1;
6941 bool BPPLR = BPPLRValue == 1;
6942 bool GCS = GCSValue == 1;
6943 bool SRA = SRAValue == 1;
6946 if (SRA && SRAALLValue == 1)
6947 SignTypeValue =
"all";
6950 if (SRA && SRABKeyValue == 1)
6951 SignKeyValue =
"b_key";
6953 for (
Function &
F : M.getFunctionList()) {
6954 if (
F.isDeclaration())
6961 if (
auto A =
F.getFnAttribute(
"sign-return-address");
6962 A.isValid() &&
"none" ==
A.getValueAsString()) {
6963 F.removeFnAttr(
"sign-return-address");
6964 F.removeFnAttr(
"sign-return-address-key");
6980 if (SRAALLValue == 1)
6982 if (SRABKeyValue == 1)
6991 if (
T->getNumOperands() != 2)
6994 if (!Tag || Tag->getString() !=
"llvm.loop.distribute.enable")
7003 if (
T->getNumOperands() < 1)
7008 if (S->getString().starts_with(
"llvm.vectorizer."))
7014 StringRef OldPrefix =
"llvm.vectorizer.";
7017 if (OldTag ==
"llvm.vectorizer.unroll")
7029 if (
T->getNumOperands() < 1)
7042 :
"llvm.loop.distribute.disable")});
7045 if (!OldTag->getString().starts_with(
"llvm.vectorizer."))
7050 Ops.reserve(
T->getNumOperands());
7052 for (
unsigned I = 1,
E =
T->getNumOperands();
I !=
E; ++
I)
7053 Ops.push_back(
T->getOperand(
I));
7070 if (
T->isDistinct()) {
7071 for (
unsigned I = 0, E =
T->getNumOperands();
I < E; ++
I) {
7083 Ops.reserve(
T->getNumOperands());
7094 if ((
T.isSPIR() || (
T.isSPIRV() && !
T.isSPIRVLogical())) &&
7095 !
DL.contains(
"-G") && !
DL.starts_with(
"G")) {
7096 return DL.empty() ? std::string(
"G1") : (
DL +
"-G1").str();
7099 if (
T.isLoongArch64() ||
T.isRISCV64()) {
7101 auto I =
DL.find(
"-n64-");
7103 return (
DL.take_front(
I) +
"-n32:64-" +
DL.drop_front(
I + 5)).str();
7108 std::string Res =
DL.str();
7111 if (!
DL.contains(
"-G") && !
DL.starts_with(
"G"))
7112 Res.append(Res.empty() ?
"G1" :
"-G1");
7120 if (!
DL.contains(
"-ni") && !
DL.starts_with(
"ni"))
7121 Res.append(
"-ni:7:8:9");
7123 if (
DL.ends_with(
"ni:7"))
7125 if (
DL.ends_with(
"ni:7:8"))
7130 if (!
DL.contains(
"-p7") && !
DL.starts_with(
"p7"))
7131 Res.append(
"-p7:160:256:256:32");
7132 if (!
DL.contains(
"-p8") && !
DL.starts_with(
"p8"))
7133 Res.append(
"-p8:128:128:128:48");
7134 constexpr StringRef OldP8(
"-p8:128:128-");
7135 if (
DL.contains(OldP8))
7136 Res.replace(Res.find(OldP8), OldP8.
size(),
"-p8:128:128:128:48-");
7137 if (!
DL.contains(
"-p9") && !
DL.starts_with(
"p9"))
7138 Res.append(
"-p9:192:256:256:32");
7142 if (!
DL.contains(
"m:e"))
7143 Res = Res.empty() ?
"m:e" :
"m:e-" + Res;
7148 if (
T.isSystemZ() && !
DL.empty()) {
7150 if (!
DL.contains(
"-S64"))
7151 return "E-S64" +
DL.drop_front(1).str();
7155 auto AddPtr32Ptr64AddrSpaces = [&
DL, &Res]() {
7158 StringRef AddrSpaces{
"-p270:32:32-p271:32:32-p272:64:64"};
7159 if (!
DL.contains(AddrSpaces)) {
7161 Regex R(
"^([Ee]-m:[a-z](-p:32:32)?)(-.*)$");
7162 if (R.match(Res, &
Groups))
7168 if (
T.isAArch64()) {
7170 if (!
DL.empty() && !
DL.contains(
"-Fn32"))
7171 Res.append(
"-Fn32");
7172 AddPtr32Ptr64AddrSpaces();
7176 if (
T.isSPARC() || (
T.isMIPS64() && !
DL.contains(
"m:m")) ||
T.isPPC64() ||
7180 std::string I64 =
"-i64:64";
7181 std::string I128 =
"-i128:128";
7183 size_t Pos = Res.find(I64);
7184 if (Pos !=
size_t(-1))
7185 Res.insert(Pos + I64.size(), I128);
7189 if (
T.isPPC() &&
T.isOSAIX() && !
DL.contains(
"f64:32:64") && !
DL.empty()) {
7190 size_t Pos = Res.find(
"-S128");
7193 Res.insert(Pos,
"-f64:32:64");
7199 AddPtr32Ptr64AddrSpaces();
7207 if (!
T.isOSIAMCU()) {
7208 std::string I128 =
"-i128:128";
7211 Regex R(
"^(e(-[mpi][^-]*)*)((-[^mpi][^-]*)*)$");
7212 if (R.match(Res, &
Groups))
7220 if (
T.isWindowsMSVCEnvironment() && !
T.isArch64Bit()) {
7222 auto I =
Ref.find(
"-f80:32-");
7224 Res = (
Ref.take_front(
I) +
"-f80:128-" +
Ref.drop_front(
I + 8)).str();
7232 Attribute A =
B.getAttribute(
"no-frame-pointer-elim");
7235 FramePointer =
A.getValueAsString() ==
"true" ?
"all" :
"none";
7236 B.removeAttribute(
"no-frame-pointer-elim");
7238 if (
B.contains(
"no-frame-pointer-elim-non-leaf")) {
7240 if (FramePointer !=
"all")
7241 FramePointer =
"non-leaf";
7242 B.removeAttribute(
"no-frame-pointer-elim-non-leaf");
7244 if (!FramePointer.
empty())
7245 B.addAttribute(
"frame-pointer", FramePointer);
7247 A =
B.getAttribute(
"null-pointer-is-valid");
7250 bool NullPointerIsValid =
A.getValueAsString() ==
"true";
7251 B.removeAttribute(
"null-pointer-is-valid");
7252 if (NullPointerIsValid)
7253 B.addAttribute(Attribute::NullPointerIsValid);
7256 A =
B.getAttribute(
"uniform-work-group-size");
7260 bool IsTrue = Val ==
"true";
7261 B.removeAttribute(
"uniform-work-group-size");
7263 B.addAttribute(
"uniform-work-group-size");
7274 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 bool isOldDistributeEnable(const MDTuple *T)
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 std::optional< StringRef > getModuleFlagNameSafely(const MDNode &Flag)
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 bool upgradePtrauthInitFiniArrays(Module &M)
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 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)
StructType * getType() const
Specialization - reduce amount of casting.
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...
reference emplace_back(ArgTypes &&... Args)
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)
bool isVoidTy() const
Return true if this is 'void'.
A Use represents the edge between a Value definition and its users.
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, bool > hasa(Y &&MD)
Check whether Metadata has a Value.
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.
@ Load
The value being inserted comes from a load (InsertElement only).
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.