[Flang][OpenMP] Optimize target updates of derived-type scalars - #4181
Closed
TIFitis wants to merge 10000 commits into
Closed
[Flang][OpenMP] Optimize target updates of derived-type scalars#4181TIFitis wants to merge 10000 commits into
TIFitis wants to merge 10000 commits into
Conversation
This fixes the neon intrinsics test, which has been failing since a recent change to introduce more conversion handling. The test had an inadvertantly repeated CIR check.
…lvm#218088) The RISC-V branch evaluator can resolve JALR when the base register state is known. In particular, jalr zero, zero, 0 evaluates to address zero even though it remains an indirect branch. Do not send indirect branches through the direct branch symbolization path. Leave them to the existing indirect branch handling instead. Add a regression test modeling a tail call to an undefined weak symbol relaxed by GNU ld.
…17640) Unfortunate the HIP runtime is hardcoding the amdgcn-amd-amdhsa-- prefix in the bundle ID checks, so force these to the legacy name regardless of the active triple. Hopefully clr will stop hardcoding these so we can drop this at some point. Co-authored-by: Claude (Claude-Opus-4.8) <noreply@anthropic.com>
…205741) This patch adds the support for partial reduction with EVL tail-folding by not creating a new VPExpressionEVLRecipes but just changing the last expressionRecipes from VPReductionRecipe to VPReductionEVLRecipe. Currently, all partial reductions will be converted to VPExpressionRecipe when construction. So this patch also supports the VPExpressionRecipe for EVL.
…lvm#218920) Require SymbolTable operations to implement SymbolOpInterface and store symbol names and visibility as inherent attributes. Add missing symbol properties/interfaces to GPU, OpenACC, OpenMP, EmitC, and Toy operations, and update affected tests and the GPU Python builder. Assisted-by: Codex
new unittest ifdef'ed for now llvm/unittests/IR/MetadataTest.cpp
When configured with `LLVM_ENABLE_RUNTIMES=compiler-rt;orc-rt`, following two CMake errors occur. ``` CMake Error at llvm-project/llvm/cmake/modules/AddLLVM.cmake:2245 (add_custom_target): add_custom_target cannot create target "check-orc-rt" because another target with the same name already exists. The existing target is a custom target created in source directory "llvm-project/compiler-rt/test/orc". See documentation for policy CMP0002 for more details. Call Stack (most recent call first): llvm-project/llvm/cmake/modules/AddLLVM.cmake:2326 (add_lit_target) llvm-project/orc-rt/test/CMakeLists.txt:24 (add_lit_testsuite) CMake Error at llvm-project/orc-rt/test/unit/CMakeLists.txt:1 (add_custom_target): add_custom_target cannot create target "OrcRTUnitTests" because another target with the same name already exists. The existing target is a custom target created in source directory "llvm-project/compiler-rt/lib/orc/tests". See documentation for policy CMP0002 for more details. ``` Both `compiler-rt/lib/orc/tests/CMakeLists.txt` and `orc-rt/test/unit/CMakeLists.txt` define targets named `OrcRTUnitTests`. Both `compiler-rt/test/orc/CMakeLists.txt` and `orc-rt/test/CMakeLists.txt` define targets named `check-orc-rt`. This commit changes target names in compiler-rt.
) This patch adds a description about the `--driver-mode=flang` flag. The behavior will be ensured by llvm#207658. --------- Co-authored-by: Tarun Prabhu <tarunprabhu@gmail.com>
This is common issue with lsan after exit(). `exit()` is no return, we can't expect that compiler will preserve pointers to allocations done by callers. Fixes new build bot report after upgrading base compiler to clang 23.1.0 https://lab.llvm.org/buildbot/#/builders/169/builds/26007 It probably has improved stack or registers re-use.
Combined subnodes were added to the trimming worklist as independent candidates, carrying their descendants' aggregated costs. Trimming such a subnode on its own deleted operands still referenced by the combined root, producing instructions that no longer dominated their uses. Skip CombinedVectorize subnodes in subtree-cost aggregation and worklist construction so only combined roots are evaluated as trim candidates. Fixes llvm#218974 Reviewers: Pull Request: llvm#219095
…in ASTReader (llvm#219053) Fixes a `std::set_difference` sorting violation in `ASTReader::checkTargetOptions` that causes a crash when building clang with `LLVM_ENABLE_EXPENSIVE_CHECKS` on libstdc++. `accumulateFeaturesAsWritten` sorts target features using a custom comparator that strips the `+`/`-` prefix (i.e. comparing `A.substr(1) < B.substr(1)`). However, `std::set_difference` was being called with the default `std::string::operator<` comparator, which does not match the sorting order because `+` (ASCII 0x2B) is less than `-` (ASCII 0x2D). For example, `["-cx16", "+sse2"]` is correctly sorted according to the custom comparator, but incorrectly sorted according to the default lexicographical comparator. This patch fixes the issue by passing the same custom comparator used for sorting to both `std::set_difference` calls. Fixes llvm#219046.
SimpleRemoteCA implements the transport-independent parts of the SimpleRemote protocol: opcode and header validation, setup-message encoding, hang-up encoding and decoding, pending-call tracking, and completion of controller calls onto the Session::ControllerAccess contract. Subclasses own wire framing and byte transport entirely -- they de-frame incoming messages and hand the header fields to handleMessage, and build outgoing ones from the Opcode set and the encode helpers. Header validation rejects fields the protocol requires to be unset: a Result may not carry a handler tag, a hang-up may not carry a sequence number or tag. It does not interpret fields the protocol requires to be present -- in particular a Call's handler tag is passed through as given, since the executor's job is to jump where the controller tells it to. A hang-up carries a serialized Error giving the reason the session is ending, matching the format SimpleRemoteEPC and SimpleRemoteEPCServer use, so that an orderly disconnect is distinguishable from a peer that vanished. A hang-up whose reason is a failure ends the session with that reason rather than reporting a plain end-of-session. SimpleRemoteCATest drives the protocol operations directly through a capture-only test double and observes the results via a Session.
Close llvm#218228 The root cause of the problem is the corresponding friend is not loaded at the point of ADL. This patch tries to fix this simply by loading the friends at the point of ADL. Note that this may be best efficient if there are a lot of friends. We just think it is rare. If it is really possible, we can change the structure of friends from a list to a name lookup table.
This patch simplifies FoldingSetIterator by defining it entirely in the header without the type-erased FoldingSetIteratorImpl. Now that FoldingSet uses an open-addressing hash table, we just have to march through the entire bucket array while skipping empty slots. It does not make sense to call out-of-line advance(). This also reduces the .text section size of a release build of bin/clang by 647 bytes (from 185,968,635 to 185,967,988 bytes). Assisted-by: Antigravity
Based on llvm#203992: Clang adds `c++2d` flag Also based on llvm@00c97cb libc++ C++26 support 1. Adds C++29 flag and macros (unlike Clang's `c++2d` uses the `c++29` flag, supported also by GCC) 2. Updates FTM tests with C++29 details 3. Updates generated files --------- Co-authored-by: Lucas Mellone <github.snugness349@passinbox.com> Co-authored-by: Hristo Hristov <zingam@outlook.com> Co-authored-by: A. Jiang <de34@live.cn>
The midend of LLVM, in CVP, will often transform the canonical pattern for a saturate trunc(smin(smax)) into trunc nsw (smin), as the range analysis proves that the smax is outside of range. This adds a fold back, converting the trunc with nsw + smin into G_TRUNC_SSAT_S. https://alive2.llvm.org/ce/z/xfPEXE Support for widening is added for truncsat nodes too, to prevent fallbacks. They can be widened in the same way as a standard trunc using buildPadVectorWithUndefElements.
…#218607) Many conversion patterns created `llvm.mlir.constant` with a value attribute whose type does not match the result type. The most common case was pairing an `index`-typed attribute with the converted index type: ``` llvm.mlir.constant(1 : index) : i64 ``` but there were also plain width and signedness mismatches, e.g. NVGPU's `makeI64Const` built `i64` constants from `i32` attributes, and the NVVM `fdiv` expansion used `ui32` attributes on `i32` values. Translation to LLVM IR ignores the attribute type and uses the result type, so the emitted IR was correct, but the attribute type is meaningless in this state and anything that reads it back sees the wrong type. Derive the attribute from the result type in every case; where the result is the converted index type, `LLVM::createIndexAttrConstant` now does this for all of its callers. In `ArithToLLVM`, retype the value attribute when the type converter maps `index` to a different integer type. Two element type changes must not be retyped: the low-precision float types that the converter maps to a same-width integer keep their float attribute, which `llvm.mlir.constant` accepts and takes the float semantics from, and a resource-backed elements attribute refers to a blob laid out for its own element type. Anything else is a malformed `arith.constant`, and the match fails rather than reinterpreting its value. `WmmaOpsToNvvm` builds its leading dimension attribute from the `i32` result, so reject a value that does not fit instead of narrowing it silently. This is in preparation for verifying that the two types agree. --------- Co-authored-by: Claude Opus 5 (1M context) <noreply@anthropic.com>
…#219072) Implements a variety of pointer and memory related ops (`G_GLOBAL_VALUE`, `G_PTR_ADD`, `G_LOAD`, `G_STORE`, etc.). Split from llvm#157161 ----- TODO (in future PR) - Move `G_PTRTOINT`, `G_INTTOPTR`, `G_PTRMASK` selection to tablegen (for all targets) - MAYBE try to shuffle `G_GLOBAL_VALUE` selection elsewhere (e.g. legalizer). Doesn't seem worth it? ----- Reapplication of llvm#206885 with some fixes and extensions.
Disabled new checks form 23 release that I found not comply with current style (yet).
…P dialect (llvm#217448) - Fixed the linker error: `undefined reference to'mlir::omp::IntegerWrapAroundAttr::get(mlir::MLIRContext*, bool)'` mentioned in mentioned in [PR#214165](llvm#214165 (comment)) by moving the `setOpenMPIntegerWrapAround(...)` helper function out of `flang/Tools/CrossToolHelpers.h` and placed it into the OpenMP dialect utilities.
This only inspects the parent instruction, so use use_instructions instead. Co-authored-by: Claude (Claude-Opus-4.8) <noreply@anthropic.com>
…#218354) This change adds support for `tcgen05.mma.decompress_b` and `tcgen05.mma.block_scale.decompress_b` MLIR Ops.
This was originally part of llvm#214132. However, that patch has some difficult to track down performance issue. I'm splitting this up to make the search easier.
Add a module pass that raises the alignment of global integer arrays (char, short, int), including multi-dimensional arrays, to an 8-byte boundary. This gives their base address a wider alignment, which is beneficial for the wide (double-word) loads and stores available on Hexagon. At -O1/-O2 the pass keeps byte and half-word arrays at their natural alignment to reduce .rodata size; full 8-byte alignment is applied at -O3. This size-reduction behavior can be disabled with -hexagon-disable-align-opt-byte-half. The pass is enabled by default and can be disabled with -hexagon-disable-global-array-align. Co-Authored by: Jyotsna Verma jverma@quicinc.com
…ets (llvm#219208) We sometimes run tests in sandboxed environments that deny all calls to `bind`. This breaks a few of our tests that e.g. use a mock GDB server or any other functionality involving sockets. This patch adds a requireSocketPermission decorator (and an equivalent utility for unittests) that check whether we are allowed to call bind. If we aren't allowed to call bind, we skip the few tests that need this functionality.
…lvm#219364) Every getVTList call serializes the list's raw bits into a FoldingSetNodeID and hashes it with xxh3, and every SDVTListNode carries an interned copy of that profile. The list itself is the key. Key VTLists on the ArrayRef the returned SDVTList already points at, and let the two-, three- and four-type overloads share the ArrayRef one. Aided by Opus 5
…lvm#219405) The loop only checks whether any user is a PHI, so iterate use_instructions() instead of the operands to query the parents. Co-authored-by: Claude (Claude-Opus-4.8) <noreply@anthropic.com>
A follow up from llvm#218652, a non-zero AVL guarantees a non-zero vl per the spec
isVLKnownLE and its getEffectiveImm helper used the VL operands to reach the MachineRegisterInfo. Pass it directly so they no longer depend on MachineOperand::getParent(). Co-authored-by: Claude (Claude-Opus-4.8) <noreply@anthropic.com>
…" (llvm#218962) This reverts commit 2be211b.
`target.process.memory-cache-line-size` is an unbounded `UInt64`, so
`settings set` accepts 0, and `MemoryCache` keeps the value in a
`uint32_t`, so it also accepts any multiple of 2^32, which truncates to
0. Every consumer then takes a remainder by 0.
`Process::ReadCStringFromMemory` divides on the first iteration of its
loop, before it touches inferior memory, so any address reproduces it.
On an x86_64 host that is a `SIGFPE` and lldb dies; on AArch64 `udiv` by
zero yields 0, so `addr % 0` evaluates to `addr` and the subtraction
underflows to a huge chunk size, and the bug hides.
```
$ lldb -b \
-o 'settings set target.process.memory-cache-line-size 0' \
-o 'target create --core linux-x86_64.core'
Floating point exception: 8
```
Under UBSan on any host, with
`lldb/test/API/tools/lldb-dap/coreFile/linux-x86_64.core`:
```
lldb/source/Target/Process.cpp:2388:40: runtime error: division by zero
#0 lldb_private::Process::ReadCStringFromMemory
#2 ProcessElfCore::GetMainExecutableModuleSpec
#3 ProcessElfCore::DoLoadCore
#4 lldb_private::Process::LoadCore
```
The truncating case crashes the same way while `settings show` reports a
value that is not 0, so nothing in the UI hints at the cause:
```
(lldb) settings set target.process.memory-cache-line-size 4294967296
(lldb) settings show target.process.memory-cache-line-size
target.process.memory-cache-line-size (unsigned) = 4294967296
```
The cache uses 0. 4294967297 truncates to 1, so the cache uses a
one-byte line.
Bound the property once instead of at each use, from 1 to `UINT32_MAX`,
the way `Debugger` bounds `term-width` and `term-height`. Rejecting
beats clamping, because `settings show` reads the stored value back and
would otherwise report a number the cache does not use:
```
(lldb) settings set target.process.memory-cache-line-size 4294967296
error: 4294967296 is out of range, valid values must be between 1 and 4294967295.
```
`OptionValueProperties::CreateLocalCopy` deep-copies the option values,
so the bounds also apply to each process's own collection, not just the
global one.
…19170) Update outdate comments around the code for picking the check BE formula. It includes the following updates: * define End = max(RHS,Start), as used later, use instead of incorrect max(End,Start) * precondition for the code use check min(RHS,Start) > Start - Stride * update remaining test to be consistent * Stride = umax(1, Stride) instead of umin, which matches the code above; umin would not avoid divide by 0. Preparation for llvm#218694 PR: llvm#219170
This change introduces a `RangeSet` TableGen intrinsic property. Existing `Range<idx, lo, hi>` keeps its current half-open semantics and continues to emit the existing range attribute. TableGen syntax: `RangeSet<idx, [[lo0, hi0], [lo1, hi1], ...]>` where `[lo, hi]` are closed ordered and non-overlapped intervals.
…llvm#219160) `DW_OP_implicit_value` stores its bytes in a `Value` whose internal type is `HostAddress`. The `DW_OP_piece` and `DW_OP_bit_piece` evaluators therefore mistook the backing-buffer address for a memory location and rejected valid full-width pieces. Use the saved implicit location-description kind to recognize this storage representation. Full-width byte pieces and zero-offset full-width bit pieces now preserve the backing bytes, while genuine host addresses and partial pieces retain their existing behavior. Add a unit test for both expressions from the issue. Fixes llvm#203224
…ion. (llvm#215347) While iterating users of a {s|z}ext's source operand in optimizeExtUses we explicitly replace the use with a trunc instruction thereby invalidating iterator of the loop. Update the loop to use make_early_inc_range to avoid invalidation.
Remove these intrinsics in favour of llvm.amdgcn.ballot: llvm.amdgcn.icmp llvm.amdgcn.fcmp
…es (#4045) The test binaries link the hotswap OBJECT libraries without linking amd_comgr, so each one has to supply `COMGR::ensureLLVMInitialized` itself. amd_comgr bakes in its own copy of LLVM and hides every internal symbol, so linking it for that one function would register AMDGPU into that copy's registry while the binary's own LLVM kept an empty one. Eleven binaries need the definition and eleven carried their own, identical down to the order of the initializer calls. State it once instead, in a translation unit compiled into each of them. It sits under `test-support` rather than in either test directory because the consumers span both: the transpile driver under `test-lit` and the hotswap unit tests under `test-unit`. An OBJECT library rather than a source listed per target, so the registration is compiled once and the same object handed to each of them.
Co-authored-by: Martin Luecke <mluecke@amd.com>
![github-advanced-security[bot]](https://avatars.githubusercontent.com/in/57789?s=60&v=4){kind=small}
There was a problem hiding this comment.
CodeQL found more than 20 potential problems in the proposed changes. Check the Files changed tab for more details.
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
20 participants
Add this suggestion to a batch that can be applied as a single commit.This suggestion is invalid because no changes were made to the code.Suggestions cannot be applied while the pull request is closed.Suggestions cannot be applied while viewing a subset of changes.Only one suggestion per line can be applied in a batch.Add this suggestion to a batch that can be applied as a single commit.Applying suggestions on deleted lines is not supported.You must change the existing code in this line in order to create a valid suggestion.Outdated suggestions cannot be applied.This suggestion has been applied or marked resolved.Suggestions cannot be applied from pending reviews.Suggestions cannot be applied on multi-line comments.Suggestions cannot be applied while the pull request is queued to merge.Suggestion cannot be applied right now. Please check back later.
Pack AMDGPU derived-type scalar updates into a single transfer and generated target region, avoiding one costly runtime transfer per component. Preserve existing lowering for unsupported clauses, pointer components, and non-AMDGPU targets.