When splitting a simm32 into LUI+ADDI(W). Subtract Lo12 from Val
to calculate Hi20. This replaces the old method of adding 0x800 to
Val. This change makes the math the reverse of how the LUI+ADDI(W)
create the immediate.

Due to the nature of WebAssembly, it's always better to keep
rotates instead of trying to optimize it. Commit 9485d983
disabled the generation of fsh for rotates, however these
tests ensure that future changes don't change the behaviour for
the Wasm backend that tends to have different optimization
requirements than other architectures. Also see:
https://github.com/llvm/llvm-project/issues/62703

Differential Revision: https://reviews.llvm.org/D152126

Implements parts of P1614R2 `operator<=>` for `queue`

Reviewed By: #libc, Mordante

Differential Revision: https://reviews.llvm.org/D146066

This patch forbids to write comment to BMIs for C++20 Named Modules.
Originally I thought this was helpful for language services like clangd.
But I found clangd don't want the BMI to contain comments actually. So
it is meaningless for C++20 Named Modules to keep such comments in
their BMI.

It is simple to enable this when someday we found we want this actually.

Similar to AArch64,AVR,PowerPC: 9e2d100e.

Similar to 49488490.

Differential Revision: https://reviews.llvm.org/D152088

Differential Revision: https://reviews.llvm.org/D152089

Reviewed By: Patryk27, jacquesguan

Differential Revision: https://reviews.llvm.org/D152129

This reverts commit 50a2341f.

This results in buildbot fail.

A CHECK() was firing when a call to the PACK intrinsic does not have a
VECTOR= argument and at least one dimension of the shape of the ARRAY=
argument could not be determined.  The CHECK was inappropriate, since
this can of course happen, such as when that argument is the result
of the SPREAD() intrinsic with non-constant DIM= or NCOPIES= arguments.
Replace with an if() statement.

Differential Revision: https://reviews.llvm.org/D152212

Since vfclass intruction will only set one single bit in the result, so if we only want to check 1 fp class, we could use vmseq to do it.

Reviewed By: craig.topper

Differential Revision: https://reviews.llvm.org/D151967

- add the `from_range_t` constructors and the related deduction guides;
- add the `push_range` member function.

(Note: this patch is split from https://reviews.llvm.org/D142335)

Reviewed By: #libc, ldionne

Differential Revision: https://reviews.llvm.org/D149829

In CMake, "TARGET" is a special keyword. But it's also an LLVM component, which
means downstreams may request "target" or "TARGET" from CMake. Quote such input
so "TARGET" is interpreted as a string rather than a keyword.

This is a followup to 75a0502f (D150884).

Fixes Meson's test suite and an issue which manifested identically to #61436
but appears to have been a slightly different problem.

Bug: https://github.com/mesonbuild/meson/issues/11642
Bug: https://github.com/llvm/llvm-project/issues/61436

Reviewed By: tstellar

Differential Revision: https://reviews.llvm.org/D152121

Produce a verifier error instead.

ea8f4b98 broke some build configurations
because it was enabled by default and some people are using a just built
libc/clang/LLVM to work on other projects where having a just built LLVM
libc in one of Clang's default include directories can make things
unusable.

Differential Revision: https://reviews.llvm.org/D152190

A previous patch added general support for printing via the RPC
interface. we should consolidate this functionality and get rid of the
old opcode that was used for simple testing.

Reviewed By: lntue

Differential Revision: https://reviews.llvm.org/D152211

Reviewed By: Peiming, manishucsd

Differential Revision: https://reviews.llvm.org/D152214

The logic and implementation follows the removal of no-op barriers. If
the fence is not making updates visible, either to the world or the
current thread, it is not needed. Said differently, the fences we remove
do not establish synchronization (happens-before) edges.
This allows us to eliminate some of the regression caused by:
  https://reviews.llvm.org/D145290

Different offsets can be handled by expansion rather than defaulting to
an unknown offset. Thus, [4,4] & [8,8] will result in [4, 12] rather
than [unknown, unknown].

If CUDA is not found this string will expand into nothing. We need to
surround it with a string otherwise it will cause build failures.

Differential Revision: https://reviews.llvm.org/D152209

The tests case is directly copied from https://reviews.llvm.org/D152179 authored by @aartbik

Reviewed By: aartbik

Differential Revision: https://reviews.llvm.org/D152204

If a combined loop has insufficient parallelism (= low trip count), we
might end up with too few teams/blocks. To counter that we can reduce
the number of threads per team we use. This patch implements a heuristic
and exposes a new environment variable to control the minimum of threads
to be employed in this case.

Issue reported by:
Felipe Cabarcas Jaramillo <cabarcas@udel.edu> (@fel-cab).

Reviewed By: tianshilei1992

Differential Revision: https://reviews.llvm.org/D152014

Derive the mustprogress attribute based on the willreturn attribute
or the fact that all callers are mustprogress.

Differential Revision: https://reviews.llvm.org/D94740

[Sanitizers][Darwin] In DlAddrSymbolizer, return only the module file name instead of the comlpete module path during symbolication.

rdar://108858834

Differential Revision: https://reviews.llvm.org/D152029

Consider mixed precision data type, i.e., F16 input lhs, F16 input rhs, F32 accumulation, and F32 output. This is typically written as F32 <= F16*F16 + F32.

During vectorization from linalg to vector for mixed precision data type (F32 <= F16*F16 + F32), linalg.matmul introduces arith.extf on input lhs and rhs operands.

"linalg.matmul"(%lhs, %rhs, %acc) ({
      ^bb0(%arg1: f16, %arg2: f16, %arg3: f32):
        %lhs_f32 = "arith.extf"(%arg1) : (f16) -> f32
        %rhs_f32 = "arith.extf"(%arg2) : (f16) -> f32
       %mul = "arith.mulf"(%lhs_f32, %rhs_f32) : (f32, f32) -> f32
        %acc = "arith.addf"(%arg3, %mul) : (f32, f32) -> f32
      "linalg.yield"(%acc) : (f32) -> ()
    })
There are backend that natively supports mixed-precision data type and does not need the arith.extf. For example, NVIDIA A100 GPU has mma.sync.aligned.*.f32.f16.f16.f32 that can support mixed-precision data type. However, the presence of arith.extf in the IR, introduces the unnecessary casting targeting F32 Tensor Cores instead of F16 Tensor Cores for NVIDIA backend. This patch adds a folding pattern to fold arith.extf into vector.contract

Differential Revision: https://reviews.llvm.org/D151918

issue#62488: Correct some syntax errors. Leave location and custom-operation-format unchanged, because I'm not sure.

Reviewed By: Mogball

Differential Revision: https://reviews.llvm.org/D149810

This patch adds the initial support required to support basic priting in
`stdio.h` via `puts` and `fputs`. This is done using the existing LLVM C
library `File` API. In this sense we can think of the RPC interface as
our system call to dump the character string to the file. We carry a
`uintptr_t` reference as our native "file descriptor" as it will be used
as an opaque reference to the host's version once functions like
`fopen` are supported.

For some unknown reason the declaration of the `StdIn` variable causes
both the AMDGPU and NVPTX backends to crash if I use the `READ` flag.
This is not used currently as we only support output now, but it needs
to be fixed

Reviewed By: sivachandra, lntue

Differential Revision: https://reviews.llvm.org/D151282

This patch adds support for the `malloc` and `free` functions. These
currently aren't implemented in-tree so we first add the interface
filies.

This patch provides the most basic support for a true `malloc` and
`free` by using the RPC interface. This is functional, but in the future
we will want to implement a more intelligent system and primarily use
the RPC interface more as a `brk()` or `sbrk()` interface only called
when absolutely necessary. We will need to design an intelligent
allocator in the future.

The semantics of these memory allocations will need to be checked. I am
somewhat iffy on the details. I've heard that HSA can allocate
asynchronously which seems to work with my tests at least. CUDA uses an
implicit synchronization scheme so we need to use an explicitly separate
stream from the one launching the kernel or the default stream. I will
need to test the NVPTX case.

I would appreciate if anyone more experienced with the implementation details
here could chime in for the HSA and CUDA cases.

Reviewed By: sivachandra

Differential Revision: https://reviews.llvm.org/D151735

Add some tests for combines to undo regressions caused by
0cfc6510.