And simplify formulas containing true/false
It's unclear to me how useful this is, it does make formulas more
conveniently self-contained now (we can usefully print them without
carrying around the "true/false" labels)

(while here, simplify !!X to X, too)

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

A common practice involves the creation of multiple `mbarrier` objects,
see an example below. This is particularly valuable in scenarios like
software pipelining for GEMM, where we need to generate multiple
barriers dynamically use and wait them in a loop.

PR improves `nvgpu.mbarrier.barrier` type into the
`nvgpu.mbarrier.group`. All `mbarrier` related Ops now uses this type.
Consequently, these Ops are now capable of managing multiple barriers
seamlessly.

Having `num_barriers = 4` helps us to locate mbarrier object(s) into
static shared memory. We could make the value dynamic that requires
dynamic shared memory it would complicate the codegen.

```
%barriers = nvgpu.mbarrier.create -> !nvgpu.mbarrier.group<3, num_barriers = 4>
nvgpu.mbarrier.init %barriers[%c0], %num_threads : !nvgpu.mbarrier.group<3, num_barriers = 4>
nvgpu.mbarrier.init %barriers[%c1], %num_threads : !nvgpu.mbarrier.group<3, num_barriers = 4>
nvgpu.mbarrier.init %barriers[%c2], %num_threads : !nvgpu.mbarrier.group<3, num_barriers = 4>
nvgpu.mbarrier.init %barriers[%c3], %num_threads : !nvgpu.mbarrier.group<3, num_barriers = 4>
...
scf.for %i = %c0 to %n step %c1 {
  nvgpu.mbarrier.try_wait %barriers[ (i % 4) ] ... 

  // ... Do work once mbarrier is ready 

  nvgpu.mbarrier.arrive.expect_tx %barriers[ (i + 3 % 4) ] ... 
}
```
We will have mbarrier usages like below: 
```
expect_tx[0]
expect_tx[1]
expect_tx[2]
Loop:
 try_wait mbarrier[0], expect_tx[3]
 try_wait mbarrier[1], expect_tx[0]
 try_wait mbarrier[2], expect_tx[1]
 try_wait mbarrier[3], expect_tx[2]
...
```

Reordering of possible strided nodes in bottom-to-top order requires
top-to-bottom reordering of the operands of such nodes, which is not
supported. Need to disable reordering of strided operands to avoid
compiler crashes.

Prior to this the command would simply crash when run on a running
process.

Of the three register commands, "info" was the only one missing these
requirements. On some level it makes sense because you're not going to
read a value or modify anything, but practically I think lldb assumes
any time you're going to access register related stuff, the process
should be paused.

I noticed this debugging with a remote gdb stub, so I've recreated that
scenario using attach in a new test case.

This reverts commit 0f864c7b due to
failures on expensive checks.

This cs_chain and cs_chain_preserve use InReg attribute to indicate
argument passed through SGPR.

Patch by: Acim Maravic

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

Unlike the per-lane mov*dup broadcast shuffles, broadcastsd/ss need port5 to splat across lanes

Found while reviewing a llvm-exegesis capture (and matches Agner + uops.info numbers) - I can't find any more easy wins from these captures so that will be it for now.

Found while reviewing a llvm-exegesis capture (and matches Agner + uops.info numbers)

The test tries to set the guard_size and stack_size of a thread to
SIZE_MAX / 4, which is a huge value in 64-bit systems but 1GB in 32-bit
ones.

We increase the size to 3 * (SIZE_MAX / 4) so it can also fail in 32-bit 
systems.

${CMAKE_CROSSCOMPILING_EMULATOR} will be used in the new rv32 buildbot
and is prepended automatically when we call add_custom_target in CMake,
except when we use a custom command.

There are two places where custom commands are used in libc, so we
explicitly add the ${CMAKE_CROSSCOMPILING_EMULATOR} variable there.
Other systems that don't use ${CMAKE_CROSSCOMPILING_EMULATOR} are
unaffected

Adjust the pipeline slightly to move ConstraintElim just before the loop
simplification pipeline. This increases the number of cases where SCEV
should can preserved in the future.

This also enables slightly more opportunities, by benefiting from
earlier CFG simplifications, which allow more conditions to be added.

Reviewed By: nikic, antoniofrighetto

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

One-Shot Bufferize no longer deallocates buffers, so `deallocationFn`
can be removed.

Note: There is a `bufferization.dealloc_tensor` op that now always
bufferizes to `memref.dealloc`. This op will be phased out soon.

Remove the yielded tensor analysis. This analysis was used to detect
cases where One-Shot Bufferize cannot deallocate buffers. Deallocation
has recently been removed from One-Shot Bufferize. Buffers are now
deallocated by the buffer deallocation pass. This analysis is no longer
needed.

Noticed while trying to workout some discrepancies between a llvm-mca and uica analysis - confirmed with a llvm-exegesis capture (and matches uops.info numbers)

This appears to be the reason behind Issue #14640 as well (heavy use of port5).

Predefined macro FUNCTION (and __FUNC__) in clang is not returning the
same string than MS for templated functions.

See https://godbolt.org/z/88n1rGs3b



For this test case MSVC is returning:
function: TestClass<class UnitTestNative>::TestClass
func: TestClass

---------

Co-authored-by: Reid Kleckner <rnk@google.com>

Adds support for SPV_KHR_bit_instructions.

It is only used whenever we don't need the whole Shader capability, which is a superset of this extension.

In libcpp, the `std::remove(const char*)` is a using decl in std
namespace `using ::remove`, which was not handled correctly in
`headerForAmbiguousStdSymbol`

This is extra testing for exp exp2 log log10 and log2 undef global isel.

The stack might be in a different address space, in which case, bitcast
does not work. We should use addrspacecast. As we do not support typed
pointer anymore, so we do not need a bitcast here anymore.

This revision adds an inliner interface to the UB dialect that allows
inlining of `ub.poison` operations.

The "register number" arrow should point to `LocationID`, not `LocIdx`.

To simplify handling PAuth in the machine outliner, introduce a
separate AArch64PointerAuth pass that is executed after both
Prologue/Epilogue Inserter and Machine Outliner passes.

After moving to AArch64PointerAuth, signLR and authenticateLR are
not used outside of their class anymore, so make them private and
simplify accordingly.

The new pass is added via AArch64PassConfig::addPostBBSections(),
so that it can change the code size before branch relaxation occurs.
AArch64BranchTargets is placed there too, so it can take into account
any PACI(A|B)SP instructions and not excessively add BTIs at the start
of functions.

Reviewed By: tmatheson

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

This fills out some extra cases for sin/cos testing for various types under
Global ISel, which seem to all do OK. The existing tests in
sincospow-vector-expansion.ll can be removed, as they are now covered
elsewhere.

This relands commit 96ea48ff.

This is a followup to #66988. The implementation there did not
account for the possibility of the catch object frame index
referrring to a fixed object, which is the case on win64.

This attribute makes the `enable_arm_streaming` pass ignore a function
(i.e. not add the enable streaming/za attributes). The main use case for
this is to prevent helper functions within tests being made streaming
functions.

Prior to this patch, `GeneralizeOuterUnitDimsUnPackOpPattern` would
assert that we cannot create a `tensor.empty` operation with dynamic
shapes.

The problem stems from the fact that we were not using the right builder
for the `tensor.empty` operation. Indeed, each dynamic dim needs to be
specified by an input variable.

Simply provide the dynamic dimensions to the `tensor.empty` builder to
fix that.

After 4a5bcbd5, switch instructions can now be handled in a
straight-forward manner by adding (ICMP_EQ, ConditionVal, CaseVal) for
te successor blocks per case.

We don't seem to have a use for the -amdgpu-keep-16-bit-reg-suffixes
option anymore. Was introduced in <https://reviews.llvm.org/D79435>.

Reviewed By: Joe_Nash, foad

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

Shorten the types used to i8 for cheaper verification and add test case
where 2 cases have the same destination, as suggested in #67061.