[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>
[llvm-profgen] Preserve zero-valued first PT_LOAD address (#212258)
FirstLoadableAddress was initialized with if (!FirstLoadableAddress), so
a valid first PT_LOAD at vaddr 0 was treated as unset and overwritten by
a later segment. Track whether the first loadable segment has been seen
instead.
Assisted by GPT-5
AMDGPU: Add missing mqsad-pk-insts to gfx13 frontend feature map (#213124)
fillAMDGCNFeatureMap omitted mqsad-pk-insts for gfx1310/gfx13-generic,
so clang wrongly rejected __builtin_amdgcn_mqsad_pk_u16_u8 on those
targets even though the backend enables the feature. Add it to the gfx13 case.
Co-authored-by: Claude (Claude-Opus-4.8)
[analyzer][NFC] Match the parameters order of getEndPath and finalizeVisitor with VisitNode (#212883)
Currently the ordering of the parameters in `getEndPath` and
`finalizeVisitor` functions does not match the declaration of the
`VisitNode` function. This PR resolves it by reordering the parameters
among these functions to be consistent.
[analyzer][NFC] Add test cases for the lifetime test suite (#212254)
During the development of the lifetime checkers I had many important
test cases saved locally and were never imported to the test files of
the checkers. This PR adds those missing test cases to the test files.
By adding these test cases it helps tracking the state of the lifetime
checkers and what improvements they need in order to have better
coverage and reduce the false positive rate.
*AI policy*: During the development I have consulted with claude sonnet
5 to list me test case scenarios it thinks are useful and then reviewed
its feedback and implemented/wrote those test cases by myself. Since
some of the test cases are edge cases I wanted to make sure the
implementation is on the right track.
japanese/ebview-gtk2: Fix crashes immediately after startup
Since japanese/ebview was removed in 2014, remove CONFLICTS.
Switch from REINPLACE_CMD to patchfile.
PR: 297153
Approved by: fluffy (mentor)
Co-authored-by: mce
security/globalprotect-openconnect: Mark DEPRECATED
2.0.0, which fixes the vulnerability, was released in February 2024.
The latest release is 2.6.4.
PR: 296823
Approved by: guru at unixarea.de (maintainer, timeout 2 weeks)
Approved by: fluffy (mentor)
[ubsan] Report specific CFI checks in UBSan summaries (#203341)
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.
[RegAlloc] Avoid split past block-prologue interference (#209704)
`leaveIntvAtTop` skips the target-specific block prologue when
determining the insertion point.
If the block prologue itself contains an instruction that interferes
with the interval, we must leave the interval before the block,
otherwise we will hit an assertion in `splitLiveThroughBlock`.
---------
Signed-off-by: Lukas Sommer <lukas.sommer at amd.com>
[Analysis] Optimise mul(const, (udiv %n, const)) during SCEV expansion (#212769)
We sometimes see SCEV expansions like this:
%lshr = lshr i64 %n, 2
%shl = shl i64 %lshr, 2
which originate from SCEV mul and udiv expressions:
mul(4, udiv(%n, 4))
that both share the same multiplicand, divisor. The instcombine pass
normally canonicalises this into
%and = and %n, -4
which is the final IR typically seen by the backend. In the case of SCEV
expansion during a pass like the loop vectoriser, instcombine does clean
this up. However, before this is cleaned up it will lead to
[5 lines not shown]
[AMDGPU][Docs] Memory Model: Refactor location-order to cover reads (#213044)
This shouldn't change anything semantically, only make the spec
(a) closer to the Vulkan spec, which includes reads in their location-ordered
relation: https://docs.vulkan.org/spec/latest/appendices/memorymodel.html#memory-model-location-ordered
(b) more uniform: this way, the may-see constraints are the same as the vanilla
LLVM constraints, only with all occurrences of "happens-before" replaced by
"location-ordered before".
cad/OrcaSlicer: Fix 32-bit build
LayerResult's second field is typed size_t, so std::numeric_limits::max
should also use size_t and not something related to coordinates for the
layer_id.
Reported by: pkg-fallout
Approved by: arrowd (co-mentor)
(cherry picked from commit b403418f4721af9ec4b2158539536e19ee564150)
[Sema][Serialization] Emit unused local typedefs in a deterministic order (#209639)
Sema::UnusedLocalTypedefNameCandidates is populated while iterating a Scope's
DeclsInScope, which is a SmallPtrSet whose iteration order depends on pointer
values and is therefore not stable across runs. The candidates are serialized
into the AST file -- both to assign declaration IDs and to write the
UNUSED_LOCAL_TYPEDEF_NAME_CANDIDATES record -- and are also used to emit the
deferred -Wunused-local-typedef warnings, so neither the emitted PCH/AST file
nor the diagnostics were reproducible. With deterministic compilation caching
this surfaces as a "cache poisoned" error, because two builds of the same PCH
produce different bytes.
Sort the candidates by source location at the point they are consumed, via
Sema::getSortedUnusedLocalTypedefNameCandidates(), so that both the diagnostics
and the serialized declarations are deterministic.
[flang][CodeGen] Replace fir.select* FIR-to-LLVM patterns with stubs that emit conversion errors (#212978)
`fir.select`, `fir.select_case`, `fir.select_rank`, and
`fir.select_type` are lowered to cf.* earlier in the pipeline
(`--fir-select-ops-conversion` and `--fir-polymorphic-op`). Their
FIR-to-LLVM conversion patterns are dead in a correct pipeline. Replace
them with a single templated stub
`SelectShouldHaveBeenConvertedStub<OP>` that emits `"'fir.<op>' op
should have already been converted"` and fails legalization, so running
`--fir-to-llvm-ir` standalone on stale IR reports a clear diagnostic
instead of "unable to legalize".
`Fir/convert-to-llvm.fir`'s six select* test blocks are removed (the
lowering no longer runs; CF-level coverage lives in
`Fir/SelectOpsConversion/`). `Fir/convert-to-llvm-invalid.fir` gains a
stub-error test per op. `Fir/Todo/select_case_with_character.fir` is
retargeted to check the equivalent diagnostic now emitted by
`--fir-select-ops-conversion`.
[6 lines not shown]
[ubsan] Test CFI summaries globally (#206434)
Always enable report_error_type=1 and print_summary=1 for CFI tests, and
update existing tests to check for cfi-bad-type. This allows us to
verify summary output comprehensively across the test suite.
This is a prerequisite for the change to the UBSan summary for CFI in
https://github.com/llvm/llvm-project/pull/203341.
Assisted-by: Automated tooling, human reviewed.
AMDGPU: Add missing mqsad-pk-insts to gfx13 frontend feature map
fillAMDGCNFeatureMap omitted mqsad-pk-insts for gfx1310/gfx13-generic, so
clang wrongly rejected __builtin_amdgcn_mqsad_pk_u16_u8 on those targets
even though the backend enables the feature. Add it to the gfx13 case.
Co-authored-by: Claude (Claude-Opus-4.8)
