Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Advance llvm-project and stablehlo. #2619

Merged
merged 13 commits into from
Dec 8, 2023

Conversation

stellaraccident
Copy link
Collaborator

@stellaraccident stellaraccident commented Dec 8, 2023

llvm-project: bbd2b08
stablehlo: ab709fe48de88c67717abfbd7ef17425eb95ddaf

These commits were chosen in order to account for an MLIR API break from llvm/llvm-project@3dbac2c which required a patch to stablehlo. We integrate a bit beyond that commit to deal with some revert/reapply cycles in the intervening range which were discovered in another downstream.

Further, it requires adaptation to the stablehlo API breaks introduced from openxla/stablehlo#1872 which are along for the ride.

Since some stablehlo builders were changed to directly take int64_t array refs, also traced that up some call stacks to eliminate some signed/unsigned mismatches that result.

Also adds a few TOSA tests to the passing set that seem to work now.

llvm-project: bbd2b08
stablehlo: ab709fe48de88c67717abfbd7ef17425eb95ddaf

These commits were chosen in order to account for an MLIR API break from llvm/llvm-project@3dbac2c which required a patch to stablehlo. We integrate a bit beyond that commit to deal with some revert/reapply cycles in the intervening range which were discovered in another downstream.

Further, it requires adaptation to the stablehlo API breaks introduced from openxla/stablehlo#1872 which are along for the ride.
@stellaraccident stellaraccident marked this pull request as ready for review December 8, 2023 02:04
Copy link
Collaborator

@qingyunqu qingyunqu left a comment

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

LGTM

@stellaraccident
Copy link
Collaborator Author

Hey @eric-k256, can you or someone more familiar with these TOSA lowerings look at what is going on here? I got a CI run that said that 5 tests were unexpectedly passing, so I added them. Then it said they failed. I might have done it wrong: the commits are above. If you fix it, feel free to land. This is otherwise ready to go.

This integrate picked up a handful of patches that look like they would be implicated:

  bbd2b08b95fe76bea138c1b03c1cd42ed3ee04df: [mlir][vector] Make `TransposeOpLowering` configurable (#73915) (Andrzej Warzyński on 2023-12-04 16:56:43 +0000)
  b935f1a7edccd9f7c1b6342be80e48693af0ce11: [bazel] Add missing dependency for 3a03da37a3c1e146ce7af1a1bbf8a2d3a0bf53df (Benjamin Kramer on 2023-12-04 17:40:35 +0100)
  86accd4e039ca60882caceab48701df4731a266c: Remove unused includes. NFC. (Benjamin Kramer on 2023-12-04 17:33:39 +0100)
  80ff67be8118d443f27595f6959d0468dfcf8ad7: [mlir][nvvm] Introduce `nvvm.fence.proxy` (#74057) (Guray Ozen on 2023-12-04 16:49:07 +0100)
  3a03da37a3c1e146ce7af1a1bbf8a2d3a0bf53df: [mlir][nvgpu] Add address space attribute converter in nvgpu-to-nvvm pass  (#74075) (Guray Ozen on 2023-12-04 16:48:39 +0100)
  6da578cec1395ee54f82b51f912fb85fdb6cdce3: [mlir] Add support for DIGlobalVariable and DIGlobalVariableExpression (#73367) (Justin Wilson on 2023-12-04 15:52:02 +0100)
  853682cc19e26964383af9e834215fae5966b140: [mlir][LLVIR] Apply ClangTidy finding. (Adrian Kuegel on 2023-12-04 11:20:58 +0000)
  8171eac23fe7756319444c2caa27216a1e9f046a: [mlir][Vector] Update patterns for flattening vector.xfer Ops (1/N) (#73522) (Andrzej Warzyński on 2023-12-04 10:21:32 +0000)
  10063c5a29b9dd5041ea7e0ab9569ed74ef54d32: [mlir][ArmSME] Move vector.print -> ArmSME lowering to VectorToArmSME (#74063) (Benjamin Maxwell on 2023-12-04 09:42:11 +0000)
  c9c1b3c37fa5d5c617068afe67afcafacd58a241: [mlir][memref] Fix an invalid dim loop motion crash (#74204) (Rik Huijzer on 2023-12-04 08:57:59 +0100)
  67f9cd46708e74aa65341b07eeefa4fe198d7a9e: [mlir][llvm] Fix verifier for const float (#74247) (Rik Huijzer on 2023-12-04 08:06:02 +0100)
  21a9c7e6e7b89b2000e2249a5305cf5d0b04ccd2: [mlir][AMDGPU] fix AMDGPU C API registration (#74255) (Maksim Levental on 2023-12-03 16:14:09 -0600)
  1c6a6ba43ba1a331177bd30194f97ce26117cf04: [bazel] Remove :CAPIGPU as a dependency of :MLIRPythonBindingsCore. (#74133) (Peter Hawkins on 2023-12-02 07:41:52 -0800)
  293c21db9381fde27cda46e5c3ff8bf8578e5399: [mlir][tosa] Improve lowering of tosa.conv2d (#74143) (Spenser Bauman on 2023-12-02 12:29:10 +0000)
  19e068b048591feb8fa66b164669c753090dfd3a: [MLIR][SCF] Handle more cases in pipelining transform (#74007) (Thomas Raoux on 2023-12-01 21:28:21 -0800)
  ccfc2d687c106ee8430fccd09e165e0aaea39081: [mlir][transform] Remove `cachedNames` expensive check (#73961) (Matthias Springer on 2023-12-02 02:35:55 +0100)
  fc74db466b0d2b87d2013d5e24be137f0d8b6f0a: [mlir][Linalg] Fix foldFillPackIntoFillOp to work for general cases (#74148) (Vivian on 2023-12-01 16:36:03 -0800)
  005c83380a907becbf5a6b4522fc43652c9536cd: [mlir][tensor] Fix ReifyResultShapes implementation for tensor.concat (#74157) (Quinn Dawkins on 2023-12-01 19:29:56 -0500)
  8206b75a1e8d2e00e9bc47ba8288dec1205fbd91: [mlir][sparse] fix crash when generate rotated convolution kernels. (#74146) (Peiming Liu on 2023-12-01 14:13:57 -0800)
  ed22bf69917479034aa4d2d42cbd9cb9d59cc0ae: [mlir][linalg] Fix weight dimension ordering in 2D grouped conv (#73855) (Felix Schneider on 2023-12-01 22:16:00 +0100)
  f310a5d2c13455f1d68f5654fa4258357bafeff6: [mlir][tensor] Add a tensor.concat operation (#72779) (Quinn Dawkins on 2023-12-01 15:05:29 -0500)
  171cac95a7eb1526f4d18bf8f654275656183ce4: [mlir][tensor] Fold padding_value away for pack ops when possible. (#74005) (Han-Chung Wang on 2023-12-01 11:12:58 -0800)
  bc802407d16f4aa0df9f32610e3b25b6a791c085: [mlir][sve][nfc] Merge the integration tests for linalg.matmul (#74059) (Andrzej Warzyński on 2023-12-01 17:39:48 +0000)
  f58fb8c209a5179f8f2e02e2a0816c9b1f1edb1b: [mlir][tosa] Fix lowering of tosa.conv2d (#73240) (Spenser Bauman on 2023-12-01 15:33:14 +0000)
  0d87e2577914a6384f4ad5952b8fa9b0d8e48da8: [mlir][tosa] Improve lowering to tosa.fully_connected (#73049) (Spenser Bauman on 2023-12-01 15:16:51 +0000)
  852f6be6967911de8ab6fe61b5a632366cc0804e: [mlir][tosa] Improve tosa-infer-shapes for ops consumed by non-TOSA operators (#72715) (Spenser Bauman on 2023-12-01 15:08:16 +0000)
  fdf84cbf87198d16fe17aed0c31989ee31051d82: [mlir][vector] Fix unit dim dropping pattern for masked writes (#74038) (Quinn Dawkins on 2023-12-01 10:01:28 -0500)
  65aab9e7222025f57c4bfc253d48c7b2ea8581da: [mlir][gpu] Generate multiple rank-specializations for tensor map cre… (#74082) (Adam Paszke on 2023-12-01 15:51:48 +0100)
  f42b7615b862bb5f77981f619f92877eb20adf54: [mlir][Vector] Add fold transpose(shape_cast) -> shape_cast (#73951) (Benjamin Maxwell on 2023-12-01 14:24:36 +0000)
  5a32014d82334c4c66c8cc7ae3ed2a489c07db07: [mlir] update linalg transform ops docs (Oleksandr "Alex" Zinenko on 2023-12-01 12:53:35 +0100)
  f7d91faa790630eca506a29faa560d6783edcbc0: [mlir][ArmSME] Add option to only enable streaming mode/ZA if required (#73931) (Benjamin Maxwell on 2023-12-01 10:39:01 +0000)
  f42ce1621f5f4129fb37c4a1af958e1d47344107: [mlir][sve][nfc] Update a test to use transform-interpreter (#73771) (Andrzej Warzyński on 2023-12-01 10:08:00 +0000)
  bb98227db19ae4d80af7a25a9423aae2aeaec61d: [libc][NFC] Remove named_pair (#73952) (Guillaume Chatelet on 2023-12-01 10:30:15 +0100)
  a224ddc9b4458b1b9cf0a758c974a554f0f17dc4: [mlir][nvvm] Introduce `cp.async.bulk.commit.group` (Guray Ozen on 2023-12-01 10:21:50 +0100)
  5a4ca51a91ff28b1d6bdde5403144c29b86e4b54: [mlir] notify insertion of parent op first when cloning (#73806) (jeanPerier on 2023-12-01 10:03:02 +0100)
  2c976a1fac5c0d6fe1cd7c3637f3d16cc378f52b: [libc] Fix _Float16 detection for x86 (#73947) (Guillaume Chatelet on 2023-12-01 09:47:26 +0100)
  1726b65e4c273d55dd54838a742b03caff4abcdd: [MLIR][Vector] Refactor tests for contract -> OP transforms (4/N) (#73807) (Andrzej Warzyński on 2023-12-01 08:45:13 +0000)
  9557fcca563dba3dd31769c297bb3b97d6e614f9: [libc] Fix lint message (#73956) (Guillaume Chatelet on 2023-12-01 09:32:22 +0100)
  6402706a00e806aef62836cd1a13a412349121eb: [mlir] Fix the link of libcuda.so in MLIRGPUTransforms to not use fully qualified path (#74018) (Mehdi Amini on 2023-11-30 19:30:05 -0800)
  b6cad75e077e1c68449b18ffeb6e93f18463464f: [mlir][sparse] refactoring: using util functions to query the index to load from position array for slice-driven loop. (#73986) (Peiming Liu on 2023-11-30 16:40:11 -0800)

@eric-k256
Copy link
Collaborator

I don't see anything wrong in what you've done. I built this PR locally, and can see the 5 tests unexpectedly pass. Switching to your commit adding them to the pass list and everything runs with no failures.

I retriggered the github checks to make sure it wasn't a weird glitch specific to that build.

@stellaraccident stellaraccident force-pushed the integrate_llvm_stablehlo_20231207 branch from cc910f7 to 5b6535b Compare December 8, 2023 04:27
@stellaraccident
Copy link
Collaborator Author

I don't see anything wrong in what you've done. I built this PR locally, and can see the 5 tests unexpectedly pass. Switching to your commit adding them to the pass list and everything runs with no failures.

I retriggered the github checks to make sure it wasn't a weird glitch specific to that build.

I think this is a difference between nightly PT and stable. Trying something else.

@stellaraccident stellaraccident merged commit 8252656 into main Dec 8, 2023
5 checks passed
@stellaraccident stellaraccident deleted the integrate_llvm_stablehlo_20231207 branch December 8, 2023 07:13
mgehre-amd pushed a commit to Xilinx/torch-mlir that referenced this pull request Feb 23, 2024
llvm-project: bbd2b08
stablehlo: ab709fe48de88c67717abfbd7ef17425eb95ddaf

These commits were chosen in order to account for an MLIR API break from
llvm/llvm-project@3dbac2c
which required a patch to stablehlo. We integrate a bit beyond that
commit to deal with some revert/reapply cycles in the intervening range
which were discovered in another downstream.

Further, it requires adaptation to the stablehlo API breaks introduced
from openxla/stablehlo#1872 which are along for
the ride.

Since some stablehlo builders were changed to directly take int64_t
array refs, also traced that up some call stacks to eliminate some
signed/unsigned mismatches that result.

Also adds a few TOSA tests to the passing set that seem to work now.
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
None yet
Projects
None yet
Development

Successfully merging this pull request may close these issues.

4 participants