[lldb][Windows] Enforce Python 3.11 (#209759)
https://github.com/llvm/llvm-project/pull/191159 recommended Python 3.11
or later. Now that llvm 23 has branched, we can enforce Python 3.11 as
the minimum requirement.
[lldb] Drop support for using the external shell
The lit external shell is going to be removed soon, so drop support for
using it to run the LLDB tests.
Reviewers: DavidSpickett, JDevlieghere, petrhosek, ilovepi
Pull Request: https://github.com/llvm/llvm-project/pull/209627
[flang][OpenMP] Avoid boxing constant-size trivial private arrays (#208315)
Part of #208086. `privatizeSymbol` (used for both OpenMP privatization
and `do concurrent` locals) currently boxes every array as a workaround
for `fir.array` not being directly `alloca`-able in the OpenMP→LLVMIR
translation. This PR carves out the common case: a constant-shape array
of trivial intrinsic elements (e.g. `real(8) :: xx(3)`) is now
privatized unboxed, as a plain `fir.array`.
[lldb][AArch64][NFC] Make watchpoint adjustment more clear (#209483)
NativeRegisterContextDBReg_arm64::AdjustWatchpoint has a way of
adjusting the range of a watchpoint but it took me a while to figure out
how it worked. I've rewritten it in a way I think is easier to follow,
with more explanatory comments.
It turns out that this adjustment results in us only using a subset of
what the hardware allows.
* You don't have to select a continuous range of bytes (1, 2, 3). You
can select (1, 4) for example.
* The address can be 4 byte aligned, in which case only 4 bits of the
byte select apply (though you can set up an equivalent watch with 8 byte
alignment and a different 8 bit BAS value).
[clang][ExprConst] Support virtual bases in C++26 (#204289)
Add support for virtual bases in constant expressions (P3533R2) to both the current constant interpreter as well as the bytecode interpreter.
[X86] combineINSERT_SUBVECTOR - enable shuffle combining of insertsub(extractsub(shuffle),extractsub(shuffle)) patterns (#209735)
Allows combineX86ShuffleChainWithExtract to more consistently fold
patterns from wider sources
[orc-rt] Fix WrapperFunctionBuffer empty range construction (#209751)
orc_rt_CreateWrapperFunctionBufferFromRange(const char *, size_t) left
the constructed buffer's data member uninitialized when the size
argument was zero. This could result in the returned buffer having a
non-null data field and a zero size field, which is a (malformed)
out-of-band error value, not an empty buffer.
Update orc_rt_CreateWrapperFunctionBufferFromRange to zero-initialize
the data field so that the size == 0 case yields a correctly formed
empty buffer.
[mlir][memref] Support truncating copied dims to non-unit in reinterpret_cast elision (#209536)
Relax the `memref.copy`through `memref.reinterpret_cast` rewrite to
support same-dimension slices where strictly one non-unit source
dimension size is truncated to a smaller non-unit result dimension.
Previously, the rewrite only supported truncating to unit size:
```mlir
memref<1xMxNxf32>
to memref<1xMx1xf32, strided<[M*N, N, 1], offset: OFF>>
```
This change relaxes constraints to also support:
```mlir
memref<1xMxNxf32>
to memref<1xMxKxf32, strided<[M*N, N, 1], offset: OFF>>
```
where `K < N`.
[7 lines not shown]
AMDGPU: Migrate GlobalISel CodeGen tests to amdgpu subarch triple (43) (#209733)
Mechanically migrate the command-line target spelling on llc/opt RUN
lines in llvm/test/CodeGen/AMDGPU from -mtriple=amdgcn ... -mcpu=<gfx> to the
folded subarch triple form (e.g. -mtriple=amdgpu9.00-amd-amdhsa), dropping the
redundant -mcpu.
Co-Authored-By: Claude <noreply at anthropic.com> (Claude-Opus-4.8)
AMDGPU: Migrate GlobalISel CodeGen tests to amdgpu subarch triple (42) (#209732)
Mechanically migrate the command-line target spelling on llc/opt RUN
lines in llvm/test/CodeGen/AMDGPU from -mtriple=amdgcn ... -mcpu=<gfx> to the
folded subarch triple form (e.g. -mtriple=amdgpu9.00-amd-amdhsa), dropping the
redundant -mcpu.
Co-Authored-By: Claude <noreply at anthropic.com> (Claude-Opus-4.8)
[CIR][AMDGPU] Add support for AMDGCN rsq and rsq_clamp builtins (#197349)
Adds codegen for the following AMDGCN reciprocal square root builtins:
- __builtin_amdgcn_rsq (double)
- __builtin_amdgcn_rsqf (float)
- __builtin_amdgcn_rsqh (half)
- __builtin_amdgcn_rsq_bf16 (bfloat16)
- __builtin_amdgcn_rsq_clamp (double)
- __builtin_amdgcn_rsq_clampf (float)
These are lowered to the corresponding `llvm.amdgcn.rsq` and
`llvm.amdgcn.rsq.clamp` intrinsic calls.
clang: Add OpenMP driver test for invalid target IDs (#209740)
Submit test that already exists in the rocm fork. This stresses
the error cases when using the legacy OpenMP -march target specifier,
which appears to be missing upstream.
clang-offload-bundler: Remove conflicting-target-ID diagnostic
This was checking that you did not try to combine arches with
any mode vs. a specific mode (e.g, gfx90a and gfx90a:xnack+). I
don't see any point in this diagnostic. All the modes have a natural
interpretation as distinct targets with a selection preference. It's
more defensible to have this rule in the user facing clang driver,
but not the low level binary utilities. This reduces the surface area
of some special case target ID parsing.
Co-authored-by: Claude (Opus 4.8)
AMDGPU: Migrate GlobalISel CodeGen tests to amdgpu subarch triple (41) (#209731)
Mechanically migrate the command-line target spelling on llc/opt RUN
lines in llvm/test/CodeGen/AMDGPU from -mtriple=amdgcn ... -mcpu=<gfx> to the
folded subarch triple form (e.g. -mtriple=amdgpu9.00-amd-amdhsa), dropping the
redundant -mcpu.
Co-Authored-By: Claude <noreply at anthropic.com> (Claude-Opus-4.8)
[AArch64] Remove HCX feature flag from backend (#209477)
This patch removes +hcx option from the llvm and removes guarding of
HCRX_EL2 system register on it.
AMDGPU: Migrate GlobalISel CodeGen tests to amdgpu subarch triple (40) (#209730)
Mechanically migrate the command-line target spelling on llc/opt RUN
lines in llvm/test/CodeGen/AMDGPU from -mtriple=amdgcn ... -mcpu=<gfx> to the
folded subarch triple form (e.g. -mtriple=amdgpu9.00-amd-amdhsa), dropping the
redundant -mcpu.
Co-Authored-By: Claude <noreply at anthropic.com> (Claude-Opus-4.8)
[RFC][AMDGPU] Add BARRIER address space
Add a new BARRIER address space that is used for global variables that are used to represent the barrier IDs in GFX12.5.
These barrier addresses just have values corresponding 1-1 to barrier IDs. They are still implemented on top of LDS, but the offsetting happens during an addrspacecast to generic, not whenever the barrier GV is used.
The motivation for this is to make the relation between LDS and barrier GVs explicit in the compiler. It does add a bit more complexity, but that complexity was already there, just hidden by pretending barrier GVs were actual LDS.
[AMDGPU] Add synthetic apertures and use them for barriers
Define what a synthetic aperture is, and adjust the barrier AS
to use this new system. This makes the barrier AS even safer to
use as now we can use all 32 bits of it without ever risking
hitting a valid address of any kind (LDS or outside LDS).
[clang][AMDGPU] Clean-up handling of named barrier type
- Allow the type in struct/classes in very limited circumstances. The goal is to enable creating trivial wrappers around the named barrier variable, but ensure we can't get into situations where things would get awkward. Currently this means we only allow the named barrier in RecordDecls with exactly 1 field, that have no base class, and are not inherited.
- Use a `amdgpu_barrier` LangAS for this type that currently maps to the local AS. This allows easy switching to the barrier AS in a future patch.
[runtimes] Add explicit offload arch tool dependencies (#206076)
Needed for the offload unittests which detect the target arch at
configure time if not forced by OFFLOAD_TESTS_FORCE_AMDGPU_ARCH. Bug had
been masked by the dependency on flang, which we recently removed in
https://github.com/llvm/llvm-project/pull/198205.
Claude assisted with this patch.
