[TableGen] Speed up intrinsic lookups with DenseMap (NFC) (#213198)
Store in DenseMap, avoids linear lookups.
Speeds up NVPTX -gen-dag-isel and -gen-instr-info by 10%.
RISCV shows similar gains.
Neutral or marginally beneficial for other targets.
AMDGPU: Add missing msad-insts to gfx13 frontend feature map
fillAMDGCNFeatureMap omitted msad-insts for gfx1310/gfx13-generic, so
clang wrongly rejected __builtin_amdgcn_msad_u8 on those targets even
though the backend enables the feature (FeatureGFX9 generation, inherited
through FeatureGFX13). Add it to the gfx13 case.
Co-authored-by: Claude (Claude-Opus-4.8)
[MachineLICM] Use `RegisterClassInfo::getRegPressureSetLimit` (#211715)
This recommits #119826, which taught `MachineLICM` to use
`RegisterClassInfo` when computing register pressure limits so
reserved registers are accounted for (#118787).
The original change was reverted by eeac0ff because it increased
compile time by causing repeated `RegisterClassInfo` computations.
This PR is based on #210826, in which `MachineRegisterClassInfo`
analysis pass was added. `MachineRegisterClassInfo` is required
by `MachineLICM` now, but the intervening machine passes that do
not affect `RegisterClassInfo` now preserve it, so the analysis
is reused instead of recomputed.
Assisted-by: TRAE CLI (GPT-5.5)
[Mips] Fix $gp was restored when used as global register variable (#201778)
The function `eliminateDeadMI` would check `if (MRI.isReserved(Reg))`,
now we only set GP to reserved when `!Subtarget.isABICalls()`. So
`eliminateDeadMI` delete the `move $gp, $4`. And we would restore $gp
after instr selection through `$gp_64 = LD $sp_64, 8`.
Check the module metadata `llvm.named.register.$28` to detect if $28 is
used as global register. Then append new conditon when set $gp to
reserverd status and return CalleeSavedRegs without $gp.
Fix #176546.
GlobalISel: Fix lowerMergeValues when dst type is float (#212795)
Lowering is done in the integer domain. Similar to the type
mismatch when dst is a pointer, make a bitcast when dst is float.
AMDGPU: Give v_cvt_sr_pk_bf16_f32 its own subtarget feature (#213150)
v_cvt_sr_pk_bf16_f32 was gated on bf16-cvt-insts, but that feature is
also present on gfx950 where the (non-sr) v_cvt_pk_bf16_f32 was first
added. The stochastic-rounding v_cvt_sr_pk_bf16_f32 was only added
for gfx1250 and has no gfx950 encoding, so it would mis-select and
later hit the "Invalid opcode" assert. Introduce
cvt-sr-pk-bf16-f32-inst, currently added to gfx13 and 125*
Co-authored-by: Claude (Claude-Opus-4.8)
[lldb] Only preload a Module's symbols once (#213094)
Creating several targets for the same file concurrently could deadlock.
Targets share Module objects. Therefore, each concurrent
`SBDebugger::CreateTarget()` call creates a `PreloadSymbols()` task for
the *same* Module. The thread pool running those tasks can pick up a
duplicate preload task for the same Module. The mutex used by the Module
is recursive, so the thread starts the task again, enters the same
`std::call_once` and deadlocks.
Preloading only needs to happen once: it does work taht a later lookup
would do anyway. If it's already in progress, skip it. This deduplicate
the task that causes the deadlock.
At desk, this fixes a timeout in
`api/multiple-targets/TestMultipleTargets.py`, roughly 1/40 runs. To
reproduce the issue, I shrunk the thread pool to 2 threads, which makes
it always timeout.
[lldb] Convert generic register names in a case insensitive way (#213001)
Part of #212778.
There are 3 types of register name:
* The primary name, displayed by default.
* An optional alias, for example AArch64 "lr" is also "x30".
* Generic convenience names like "sp", "ra" and so on.
The first two types are handled case insensitively, but generic names
were not. For example:
(lldb) register read RA
error: Invalid register name 'RA'
(lldb) register read ra
lr = 0x0000fffff7e27400
In this change I've fixed that.
This does close a workaround for #212778, where you could get to the
[2 lines not shown]
[SystemZ][z/OS] Implement emitCommonSymbol in MCGOFFStreamer (#210179)
Common globals were not emitted in GOFF object files due to a missing
emitCommonSymbol implementation. This adds the implementation to emit
the required SD/ED/PR ESD records in the GOFF object file.
(cherry picked from commit 705d0ec6115cef3232eafe6221f42f78e7b884e8)
[DTLTO] Overlap temporary file removal (#209423)
Deleting the temporary files produced by the DTLTO pipeline can be
expensive on Windows hosts. For a Clang link (Debug build with
sanitizers and instrumentation) using an optimized toolchain (PGO
non-LTO, llvmorg-22.1.0) on a Windows 11 Pro (Build 26200), AMD Family
25 @ ~4.5 GHz, 16 cores/32 threads, 64 GB RAM machine, the mean duration
of the "Remove DTLTO temporary files" time trace scope was 1267.789 ms
(measured over 10 runs).
This patch performs the deletions on a background thread, allowing them
to overlap with the tail of the link to hide this cost.
This is a re-implementation of the asynchronous cleanup idea from
https://github.com/llvm/llvm-project/pull/186988, which had to be
reverted in https://github.com/llvm/llvm-project/pull/189043 because
cleanup was not guaranteed to complete before LLD invoked
timeTraceProfilerCleanup(). In certain cases timeTraceProfilerCleanup()
was called before temporary file deletion had completed in LLD, which
[11 lines not shown]
GlobalISel: Don't use GISelMatchGenericTypes for floating point opcodes (#212814)
With GISelMatchGenericTypes, AMDGPU had the same input pattern for
G_ATOMICRMW_FADD that checks for v2s16, so the one that is first
in the tablegen table always fires (it was v2f16 in this case).
Should explicitly check for v2f16/v2bf16.
[LoopIdiom] Update MemorySSA when optimizing CRC with lookup table (#213054)
Although the lookup table optimization of CRC loops inserts a `load`
instruction, MemorySSA is not properly updated, and is oblivious to the
new memory access. Insert a memory use immediately after creating the
`load` instruction, and verify MemorySSA at the end of the optimization
if applicable.
[InlineSpiller][AMDGPU] Implement subreg reload during RA spill
Currently, when a virtual register is partially used, the
entire tuple is restored from the spilled location, even if
only a subset of its sub-registers is needed. This patch
introduces support for partial reloads by analyzing actual
register usage and restoring only the required sub-registers.
This improvement enhances register allocation efficiency,
particularly for cases involving tuple virtual registers.
For AMDGPU, this change brings considerable improvements
in workloads that involve matrix operations, large vectors,
and complex control flows.
[AMDGPU] Test precommit for subreg reload
This test currently fails due to insufficient
registers during allocation. Once the subreg
reload is implemented, it will begin to pass
as the partial reload help mitigate register
pressure.
[CodeGen] Enhance createFrom for sub-reg aware cloning
Instead of just cloning the virtual register, this
function now creates a new virtual register derived
from a subregister class of the original value.
[ubsan] Report specific CFI checks in UBSan summaries (#213224)
Instead of reporting generic `cfi-bad-type` in UBSan summaries for CFI
failures, report the specific CFI check kind (e.g., `cfi-vcall`,
`cfi-nvcall`, `cfi-icall`, `cfi-mfcall`, `cfi-derived-cast`,
`cfi-unrelated-cast`).
This is done by splitting CFIBadType into specific error types in
ubsan_checks.inc, and updating the handlers to report the appropriate
ErrorType based on the check kind. The suppression flag name for all of
them remains `cfi` to maintain backward compatibility.
Also replaces cfi-bad-type expectations in the existing tests.
Assisted-by: Automated tooling, human reviewed.
[RISCV] Deduplicate frame base address materializations for large stack frames (#211748)
For large stack frames, `eliminateFrameIndex` materializes the frame
base with a `LUI+ADD` sequence for every frame index access. When
multiple accesses share the same 4096-aligned base, this generates
redundant `LUI+ADD` sequences that `MachineLateInstrsCleanup` cannot
eliminate because it operates on individual instructions.
Introduce `PseudoAddUpperImm` to represent the `LUI+ADD` sequence as a
single pseudo instruction during frame index elimination. This enables
`MachineLateInstrsCleanup` to CSE identical frame base computations
within a basic block. The pseudo is expanded back to `LUI+ADD` in
`RISCVPostRAExpandPseudo`.
This patch improves SPEC CPU2026 `cactus` performance by 18% on the
SpaceMit K3 processor.
### Example
[18 lines not shown]
[mlir][NVVM] Add memory clobber support to inline_ptx and BasicPtxBui… (#212937)
PTX with memory side effects (stores, atomics, mbarrier operations with
acquire/release semantics) emitted through `nvvm.inline_ptx` lowers to
`llvm.inline_asm` with register constraints only. Without a `~{memory}`
clobber, LLVM may reorder memory accesses across the inline assembly.
There was no way to express the clobber through this op or the
`BasicPtxBuilderInterface` machinery.
This patch adds, as discussed with @grypp:
- A `hasMemoryClobber` interface method (default `false`) on
`BasicPtxBuilderOpInterface`; when it returns `true`, `PtxBuilder`
appends `~{memory}` to the constraints of the generated inline assembly,
after all register constraints and tied indices.
- An opt-in `memory_clobber` boolean attribute (default `false`) on
`nvvm.inline_ptx` exposing this. Existing behavior is unchanged unless
the attribute is set to `true`.
For example:
[9 lines not shown]
[mlir][tosa] Switch matmul_t output to bf16 from fp16 (#212970)
Note that this implements the TOSA specification change:
https://github.com/arm/tosa-specification/pull/67
---------
Signed-off-by: Jeremy Johnson <jeremy.johnson at arm.com>
