LLVM/project 5612c27llvm/include/llvm/IR Intrinsics.td, llvm/test/Transforms/GVNSink pseudo-probe.ll

[IR] Mark llvm.pseudoprobe arguments as immarg (#224669)

Per LangRef, llvm.pseudoprobe requires constant integer arguments.
Enforce this by marking them as immarg.

In particular, this prevents invalid sinking of llvm.pseudprobe calls in
GVNSink, which would introduce a phi node for a pseudoprobe argument.

Fixes #166306.
DeltaFile
+46-0llvm/test/Transforms/GVNSink/pseudo-probe.ll
+3-1llvm/include/llvm/IR/Intrinsics.td
+1-1llvm/test/Transforms/SampleProfile/pseudo-probe-emit.ll
+1-1llvm/test/Transforms/SampleProfile/pseudo-probe-emit-macho.ll
+51-34 files

LLVM/project b9b829eflang/docs ModFiles.md, flang/include/flang/Support Fortran-features.h

[flang][cuda] Restrict the MANAGED(IMPLICIT) spelling to module files
DeltaFile
+22-1flang/docs/ModFiles.md
+14-0flang/test/Semantics/CUDA/cuda-managed-implicit-modfile.cuf
+4-3flang/include/flang/Support/Fortran-features.h
+3-2flang/lib/Parser/Fortran-parsers.cpp
+4-0flang/lib/Semantics/mod-file.cpp
+3-0flang/lib/Support/Fortran-features.cpp
+50-66 files

LLVM/project 8f0baabllvm/lib/ExecutionEngine/Orc COFFPlatform.cpp

[ORC] Drive COFFPlatform bootstrap calls via proxies (#225007)

Replace the three callSPSWrapper sites in bootstrapCOFFRuntime (platform
bootstrap, register_jitdylib, register_object_sections) with Proxy
dispatch. CI descriptors are kept local to the .cpp, since these are
COFFPlatform specific.
DeltaFile
+50-8llvm/lib/ExecutionEngine/Orc/COFFPlatform.cpp
+50-81 files

LLVM/project fe62ed5llvm/lib/Target/NVPTX NVPTXISelLowering.cpp, llvm/test/CodeGen/NVPTX wmma-ptx94-sm120f.py wmma-ptx94-sm90a.py

[NVPTX] Add support for ldmatrix extensions introduced in PTX 9.4 (#224264)

This patch adds support for `.s8.s4` types for `ldmatrix` instruction 
for `.m8n16` shape introduced in PTX 9.4.

PTX ISA Reference:
https://docs.nvidia.com/cuda/parallel-thread-execution/index.html#warp-level-matrix-instructions-ldmatrix

---------

Signed-off-by: DharuniRAcharya <dharunira at nvidia.com>
DeltaFile
+33-1llvm/test/CodeGen/NVPTX/wmma.py
+14-0llvm/test/CodeGen/NVPTX/wmma-ptx94-sm120f.py
+14-0llvm/test/CodeGen/NVPTX/wmma-ptx94-sm90a.py
+14-0llvm/test/CodeGen/NVPTX/wmma-ptx94-sm110f.py
+14-0llvm/test/CodeGen/NVPTX/wmma-ptx94-sm100f.py
+6-3llvm/lib/Target/NVPTX/NVPTXISelLowering.cpp
+95-43 files not shown
+105-69 files

LLVM/project 1784573mlir/lib/Target/LLVMIR LoopAnnotationTranslation.cpp LoopAnnotationImporter.cpp, mlir/test/Target/LLVMIR loop-metadata.mlir

[mlir][LLVMIR] Fix loop follow-up metadata import and export (#223376)

## Problem

MLIR currently exports loop follow-up metadata by wrapping its
properties in an additional self-referencing loop ID:

```llvm
!1 = !{!"llvm.loop.unroll.followup_unrolled", !2}
!2 = distinct !{!2, !3}
!3 = !{!"llvm.loop.mustprogress"}
```

LLVM expects the properties directly after the follow-up name:

```llvm
!1 = !{!"llvm.loop.unroll.followup_unrolled", !3}
!3 = !{!"llvm.loop.mustprogress"}
```

    [42 lines not shown]
DeltaFile
+65-38mlir/lib/Target/LLVMIR/LoopAnnotationImporter.cpp
+83-9mlir/test/Target/LLVMIR/Import/metadata-loop.ll
+37-1mlir/test/Target/LLVMIR/Import/import-failure.ll
+15-19mlir/test/Target/LLVMIR/loop-metadata.mlir
+7-3mlir/lib/Target/LLVMIR/LoopAnnotationTranslation.cpp
+207-705 files

LLVM/project 01e5735clang-tools-extra/clangd Diagnostics.cpp Diagnostics.h, clang-tools-extra/clangd/unittests FeatureModulesTests.cpp

[clangd] Extend FeatureModule hooks (#221054)

Feature modules may need to participate at points in AST construction
that `beforeExecute()` and `sawDiagnostic()` cannot represent.

This PR adds:
- `beforePPCallbacks()` for installing preprocessing observers before
clangd starts collecting include and macro events.
- `afterExecute()` for work that needs a completed AST after token
collection and traversal-scope restriction.
- `finalizeDiagnostic()` for transformations that need the complete
diagnostic, including its notes and fixes.

For example, clang-tidy needs to register preprocessing callbacks before
preamble events are replayed, run AST matchers after clangd restricts
the traversal scope, and process diagnostics after their notes and fixes
are attached.

These changes prepare moving the clang-tidy implementation into a

    [2 lines not shown]
DeltaFile
+131-11clang-tools-extra/clangd/unittests/FeatureModulesTests.cpp
+19-0clang-tools-extra/clangd/FeatureModule.h
+13-0clang-tools-extra/clangd/Preamble.cpp
+10-0clang-tools-extra/clangd/ParsedAST.cpp
+6-2clang-tools-extra/clangd/Diagnostics.h
+3-0clang-tools-extra/clangd/Diagnostics.cpp
+182-136 files

LLVM/project 7d8c2a2clang/lib/CodeGen CGCall.cpp CGExprCXX.cpp, clang/test/CodeGenCXX new_hot_cold.cpp operator-new.cpp

[Clang] Do not apply inaccessiblememonly to ::operator new() (#224316)

Do not apply `memory(inaccessiblemem: readwrite, errnomem: write)` to
`::operator new()` style calls (as opposed to `new T`). This matches the
cases where we mark the call as `builtin`, i.e. we only consider it
inaccessiblememonly in the cases where the allocation is also elidable.

I've left the old `noalias` handling alone, so that part keeps being
applied to `::operator new()` as before. Let me know if I should move
that to EmitNewDeleteCall() as well.
DeltaFile
+13-3clang/lib/CodeGen/CGExprCXX.cpp
+2-9clang/lib/CodeGen/CGCall.cpp
+9-0clang/test/CodeGenCXX/operator-new.cpp
+2-2clang/test/CodeGenCXX/new_hot_cold.cpp
+26-144 files

LLVM/project 0425600llvm/test/CodeGen/AArch64 scalarize-vector-load.ll itofp-bf16.ll, llvm/test/CodeGen/PowerPC v2i64_scalar_to_vector_shuffle.ll

DAG: Handle load in SimplifyDemandedVectorElts

This improves some AMDGPU cases and avoids future regressions.
The combiner likes to form shuffles for cases where an extract_vector_elt
would do perfectly well, and this recovers some of the regressions from
losing load narrowing.

AMDGPU, Arch64 and RISCV test changes look broadly better. Other targets have
some improvements, but mostly regressions. In particular X86 looks much
worse. I'm guessing this is because it's shouldReduceLoadWidth is wrong.

I mostly just regenerated the checks. I assume some set of them should
switch to use volatile loads to defeat the optimization.
DeltaFile
+1,015-978llvm/test/CodeGen/X86/ucmp.ll
+406-344llvm/test/CodeGen/X86/avx512-shuffles/partial_permute.ll
+276-276llvm/test/CodeGen/AArch64/itofp-bf16.ll
+147-107llvm/test/CodeGen/AArch64/scalarize-vector-load.ll
+165-74llvm/test/CodeGen/RISCV/rvv/vector-deinterleave-fixed.ll
+111-101llvm/test/CodeGen/PowerPC/v2i64_scalar_to_vector_shuffle.ll
+2,120-1,880163 files not shown
+4,580-4,160169 files

LLVM/project 3eaddb5llvm/lib/Target/SystemZ SystemZTargetMachine.cpp CMakeLists.txt, llvm/test/CodeGen/SystemZ la-01.ll

[SystemZ] Try to align globals to at least 2 bytes (#224586)

s390x has a minimum ABI alignment of 2 for globals. However, private
globals introduced by optimizations (including directly in the frontend
in the case of promotion of large const variables) are not required to
respect this ABI alignment. To avoid having to go through GOT to load an
unaligned addressed, we should try to raise the alignment of such
globals in the backend. This is done through a new SystemZ-specific
module pass.
DeltaFile
+61-0llvm/lib/Target/SystemZ/SystemZAlignGlobals.cpp
+13-2llvm/test/CodeGen/SystemZ/la-01.ll
+3-0llvm/lib/Target/SystemZ/SystemZ.h
+1-0llvm/lib/Target/SystemZ/SystemZTargetMachine.cpp
+1-0llvm/lib/Target/SystemZ/CMakeLists.txt
+79-25 files

LLVM/project ea52030llvm/lib/Target/RISCV/AsmParser RISCVAsmParser.cpp, llvm/test/MC/RISCV/corev XCVbitmanip.s XCVbitmanip-invalid.s

[RISCV][MC] Reject invalid combined immediates for `cv.insert` (#224985)

The CV32E40P specification requires `Is3 + Is2 < 32` for `cv.insert`.
This commit fixes the MC part problem in LLVM.

Link:
https://github.com/openhwfoundation/cv32e40p/blob/cv32e40p_v1.8.3/docs/source/instruction_set_extensions.rst#L600-L604
DeltaFile
+6-6llvm/test/MC/RISCV/corev/XCVbitmanip.s
+12-0llvm/test/MC/RISCV/corev/XCVbitmanip-invalid.s
+5-0llvm/lib/Target/RISCV/AsmParser/RISCVAsmParser.cpp
+23-63 files

LLVM/project f35003cclang/test/CodeGenCXX builtin-clear-padding-codegen.cpp, llvm/lib/Target/Xtensa XtensaS3DSPInstrInfo.td

Merge branch 'main' into users/adams381/cir-callconv-nonbyval-param-noload
DeltaFile
+11,213-152llvm/test/CodeGen/RISCV/rvv/vector-interleave-fixed.ll
+4,377-4,141llvm/test/CodeGen/AMDGPU/amdgcn.bitcast.1024bit.ll
+5,753-0llvm/lib/Target/Xtensa/XtensaS3DSPInstrInfo.td
+2,822-2,774llvm/test/CodeGen/AMDGPU/bf16.ll
+4,402-1,088clang/test/CodeGenCXX/builtin-clear-padding-codegen.cpp
+2,633-2,496llvm/test/CodeGen/AMDGPU/amdgcn.bitcast.512bit.ll
+31,200-10,6514,441 files not shown
+221,115-97,8444,447 files

LLVM/project c64686fclang/lib/CodeGen CGStmtOpenMP.cpp, clang/lib/Sema SemaOpenMP.cpp

[Clang][OpenMP][NFC] Use `getLimitedValue` for loop transform counts  (#224940)

Use `getLimitedValue()` when reading integer counts for `unroll, split,
interchange, fuse` transforms.

It avoids direct `getZExtValue()` calls and follows the existing OpenMP
count handling style.
DeltaFile
+5-5clang/lib/Sema/SemaOpenMP.cpp
+3-2clang/lib/CodeGen/CGStmtOpenMP.cpp
+8-72 files

LLVM/project 32150a8clang/include/clang/AST OpenMPClause.h, clang/lib/AST OpenMPClause.cpp

[Clang][OpenMP] NFC: Eliminate `Create/CreateEmpty` for scalar clause (#224935)

Use a normal constructor when the clause has a fixed size.
Keep `Create/CreateEmpty` only when the object needs extra tail storage.

Covers: `looprange, full, partial, bind, depobj, align`
DeltaFile
+32-88clang/include/clang/AST/OpenMPClause.h
+0-82clang/lib/AST/OpenMPClause.cpp
+11-11clang/lib/Sema/SemaOpenMP.cpp
+5-5clang/lib/Serialization/ASTReader.cpp
+48-1864 files

LLVM/project 6fc11f1clang/lib/CodeGen CGCall.cpp

fix ci failure
DeltaFile
+8-0clang/lib/CodeGen/CGCall.cpp
+8-01 files

LLVM/project f86d33dllvm/include/llvm/ABI FunctionInfo.h, llvm/unittests/ABI FunctionInfoTest.cpp

[ABI] Add IndirectAliased kind to abi::ArgInfo
DeltaFile
+44-0llvm/unittests/ABI/FunctionInfoTest.cpp
+20-3llvm/include/llvm/ABI/FunctionInfo.h
+64-32 files

LLVM/project 8ba8323llvm/lib/ABI TargetInfo.cpp, llvm/unittests/ABI TargetInfoTest.cpp

update
DeltaFile
+4-8llvm/unittests/ABI/TargetInfoTest.cpp
+3-3llvm/lib/ABI/TargetInfo.cpp
+7-112 files

LLVM/project 64a586fclang/lib/CIR/Dialect/Transforms/TargetLowering CIRABIRewriteContext.cpp, clang/test/CIR/CodeGen call-conv-lowering-x86_64-non-byval-thunk.cpp call-conv-lowering-x86_64-non-byval-thunk-mixed.cpp

[CIR] Read a non-byval parameter back out of its spill slot

CallConvLowering recognizes a forwarded non-byval indirect argument by the slot
its operand was loaded from.  At -O1 and above cir-simplify folds that load
away when the slot is a constant alloca, which is what CIRGen emits for a
const-qualified by-value parameter.  The walk that gives each such parameter's
slot the alignment the ABI promises now also reads the record back out of the
slot at the spill and points the parameter's call-argument uses at that load,
so an Expand, byval or coerced argument reads it too.

Assisted-by: Cursor / claude-opus-5
DeltaFile
+655-0clang/test/CIR/Transforms/abi-lowering/indirect-non-byval-forward-param.cir
+309-0clang/test/CIR/Transforms/abi-lowering/indirect-non-byval-nyi.cir
+147-12clang/lib/CIR/Dialect/Transforms/TargetLowering/CIRABIRewriteContext.cpp
+70-0clang/test/CIR/CodeGen/call-conv-lowering-x86_64-non-byval-thunk-sret.cpp
+58-0clang/test/CIR/CodeGen/call-conv-lowering-x86_64-non-byval-thunk-mixed.cpp
+50-0clang/test/CIR/CodeGen/call-conv-lowering-x86_64-non-byval-thunk.cpp
+1,289-122 files not shown
+1,329-368 files

LLVM/project 9ee03c6llvm/include/llvm/ABI TargetInfo.h, llvm/lib/ABI TargetInfo.cpp

[ABI] Add default argument/return classifiers to TargetInfo
DeltaFile
+171-0llvm/unittests/ABI/TargetInfoTest.cpp
+39-0llvm/lib/ABI/TargetInfo.cpp
+8-0llvm/include/llvm/ABI/TargetInfo.h
+1-0llvm/utils/gn/secondary/llvm/unittests/ABI/BUILD.gn
+1-0llvm/unittests/ABI/CMakeLists.txt
+220-05 files

LLVM/project 5a50c15llvm/docs CommandLine.md, llvm/include/llvm/Support CommandLine.h

[Support] Remove cl::Sink (#224953)

Nothing uses it: the last users, clang's offloading linker wrappers,
moved to OptTable. The feature would block migration to TableGen based
representation. An unknown option is now always an error.

LLM-aided
DeltaFile
+2-22llvm/lib/Support/CommandLine.cpp
+3-7llvm/unittests/Support/CommandLineTest.cpp
+0-8llvm/docs/CommandLine.md
+2-4mlir/include/mlir/Pass/PassOptions.h
+0-3llvm/include/llvm/Support/CommandLine.h
+7-445 files

LLVM/project 101bff3llvm/lib/Target/AMDGPU AMDGPUInstCombineIntrinsic.cpp, llvm/test/Transforms/InstCombine/AMDGPU llvm.amdgcn.dot.ll

[AMDGPU][InstCombine] Fold zero dot operands to accumulator

Fold AMDGPU dot intrinsics when either operand is zero.

`dot(a, 0) = 0` and `dot(0, b) = 0`, so replace the intrinsic with its accumulator.
This avoids unrelated clamp and add/sub reassociation cases.
DeltaFile
+12-24llvm/test/Transforms/InstCombine/AMDGPU/llvm.amdgcn.dot.ll
+4-0llvm/lib/Target/AMDGPU/AMDGPUInstCombineIntrinsic.cpp
+16-242 files

LLVM/project 2d7088allvm/lib/Target/AMDGPU AMDGPUInstCombineIntrinsic.cpp, llvm/test/Transforms/InstCombine/AMDGPU llvm.amdgcn.dot.ll

[AMDGPU][InstCombine] Fold constant add/sub into the dot accumulator

`amdgcn.{s,u}dot{2,4,8}(a, b, C) +/- K -> dot(a, b, C +/- K)` when both the
accumulator C and K are constants. The new constant is computed with
wrapping APInt arithmetic to match the non-clamping accumulate.

Only applies when clamp is false (the saturating accumulate does not
reassociate) and the dot has a single use. K - dot is left alone since it
would need the dot product negated.
DeltaFile
+33-66llvm/test/Transforms/InstCombine/AMDGPU/llvm.amdgcn.dot.ll
+38-0llvm/lib/Target/AMDGPU/AMDGPUInstCombineIntrinsic.cpp
+71-662 files

LLVM/project b4bfc5bllvm/test/Transforms/InstCombine/AMDGPU llvm.amdgcn.dot.ll

[AMDGPU][InstCombine] Add dot zero/constant folding tests. NFC
DeltaFile
+556-0llvm/test/Transforms/InstCombine/AMDGPU/llvm.amdgcn.dot.ll
+556-01 files

LLVM/project 90a20bbllvm/unittests/ABI CMakeLists.txt X86TargetInfoTest.cpp, llvm/utils/gn/secondary/llvm/unittests/ABI BUILD.gn

move tests to TargetInfoTest.cpp
DeltaFile
+147-0llvm/unittests/ABI/TargetInfoTest.cpp
+0-87llvm/unittests/ABI/X86TargetInfoTest.cpp
+1-0llvm/utils/gn/secondary/llvm/unittests/ABI/BUILD.gn
+1-0llvm/unittests/ABI/CMakeLists.txt
+149-874 files

LLVM/project f11d157llvm/include/llvm/ABI TargetInfo.h, llvm/lib/ABI TargetInfo.cpp

[ABI] Hoist isSingleElementStruct into TargetInfo and fix array reduction
DeltaFile
+87-0llvm/unittests/ABI/X86TargetInfoTest.cpp
+57-0llvm/lib/ABI/TargetInfo.cpp
+0-55llvm/lib/ABI/Targets/X86.cpp
+3-0llvm/include/llvm/ABI/TargetInfo.h
+147-554 files

LLVM/project a3d0da5llvm/lib/Target/X86 X86ISelLowering.cpp, llvm/test/CodeGen/X86 cmov-unpredictable-mem.mir cmov-unpredictable-flag.mir

[X86] Preserve node flags when combineCMov rebuilds a CMOV (#223469)
DeltaFile
+108-0llvm/test/CodeGen/X86/cmov-unpredictable-flag.mir
+63-0llvm/test/CodeGen/X86/cmov-unpredictable-mem.mir
+8-6llvm/lib/Target/X86/X86ISelLowering.cpp
+179-63 files

LLVM/project 606daf5llvm/lib/Target/AMDGPU AMDGPUInstCombineIntrinsic.cpp, llvm/test/Transforms/InstCombine/AMDGPU llvm.amdgcn.dot.ll

[AMDGPU][InstCombine] Fold zero dot operands to accumulator

Fold AMDGPU dot intrinsics when either operand is zero.

`dot(a, 0) = 0` and `dot(0, b) = 0`, so replace the intrinsic with its accumulator.
This avoids unrelated clamp and add/sub reassociation cases.
DeltaFile
+12-24llvm/test/Transforms/InstCombine/AMDGPU/llvm.amdgcn.dot.ll
+4-0llvm/lib/Target/AMDGPU/AMDGPUInstCombineIntrinsic.cpp
+16-242 files

LLVM/project 891a447llvm/include/llvm/ABI FunctionInfo.h, llvm/unittests/ABI FunctionInfoTest.cpp

[ABI] Add IndirectAliased kind to abi::ArgInfo
DeltaFile
+44-0llvm/unittests/ABI/FunctionInfoTest.cpp
+20-3llvm/include/llvm/ABI/FunctionInfo.h
+64-32 files

LLVM/project c007604llvm/lib/Target/AMDGPU AMDGPUInstCombineIntrinsic.cpp, llvm/test/Transforms/InstCombine/AMDGPU llvm.amdgcn.dot.ll

[AMDGPU][InstCombine] Fold constant add/sub into the dot accumulator

`amdgcn.{s,u}dot{2,4,8}(a, b, C) +/- K -> dot(a, b, C +/- K)` when both the
accumulator C and K are constants. The new constant is computed with
wrapping APInt arithmetic to match the non-clamping accumulate.

Only applies when clamp is false (the saturating accumulate does not
reassociate) and the dot has a single use. K - dot is left alone since it
would need the dot product negated.
DeltaFile
+33-66llvm/test/Transforms/InstCombine/AMDGPU/llvm.amdgcn.dot.ll
+39-0llvm/lib/Target/AMDGPU/AMDGPUInstCombineIntrinsic.cpp
+72-662 files

LLVM/project f42c0b8llvm/test/Transforms/InstCombine/AMDGPU llvm.amdgcn.dot.ll

[AMDGPU][InstCombine] Add dot zero/constant folding tests. NFC
DeltaFile
+556-0llvm/test/Transforms/InstCombine/AMDGPU/llvm.amdgcn.dot.ll
+556-01 files

LLVM/project 2855e82flang/include/flang/Parser parse-tree.h, flang/lib/Parser Fortran-parsers.cpp

[flang][cuda] Record implicit managed attribution in module files

An attribute the compiler applied under -gpu=mem:managed is written into the
module file the same way a user-written one is, so a reader cannot tell them
apart. It then treats the attribute as a user requirement: allocating such a
component in a DEVICE object is rejected, and the memory space the user did
ask for on the object no longer wins.

Spell the distinction in the module file as MANAGED(IMPLICIT), modelled on
INTENT(IN): CUDA-data-attr gains an optional parenthesized qualifier, carried
by a new CUDADataAttrSpec parse-tree node in AttrSpec and ComponentAttrSpec.
ATTRIBUTES(...) keeps the bare attribute, so the qualifier cannot be written
there.

The attribute itself is still written out, so a component keeps the same
memory space no matter which options a consumer is compiled with.

Also stop an implicitly applied attribute from making a module a definer of
CUDA symbols. Without this, adding -gpu=mem:managed to a module's build
rejects its OpenACC-only consumers over an attribute the user never wrote.
DeltaFile
+76-0flang/test/Semantics/CUDA/cuda-managed-implicit-modfile.cuf
+29-9flang/lib/Semantics/resolve-names.cpp
+19-3flang/lib/Semantics/mod-file.cpp
+13-2flang/include/flang/Parser/parse-tree.h
+10-2flang/lib/Parser/Fortran-parsers.cpp
+4-2flang/test/Parser/cuf-sanity-tree.CUF
+151-182 files not shown
+159-188 files