[ROCm] Fix Int_mm() Integration with hipblasLT #122431

petrex · 2024-03-21T20:12:56Z

The PR

fixes int_mm() /int8_gemm() integration with hipblasLT backend (require ROCm 6.0).
enables/fixes the following tests on Rocm
- test__int_mm_k_16_n_16_use_transpose_a_False_use_transpose_b_False_cuda
- test__int_mm_k_16_n_16_use_transpose_a_False_use_transpose_b_True_cuda
- test__int_mm_k_16_n_16_use_transpose_a_True_use_transpose_b_False_cuda
- test__int_mm_k_16_n_16_use_transpose_a_True_use_transpose_b_True_cuda
- test__int_mm_k_16_n_32_use_transpose_a_False_use_transpose_b_False_cuda
- test__int_mm_k_16_n_32_use_transpose_a_False_use_transpose_b_True_cuda
- test__int_mm_k_16_n_32_use_transpose_a_True_use_transpose_b_False_cuda
- test__int_mm_k_16_n_32_use_transpose_a_True_use_transpose_b_True_cuda
- test__int_mm_k_32_n_16_use_transpose_a_False_use_transpose_b_False_cuda
- test__int_mm_k_32_n_16_use_transpose_a_False_use_transpose_b_True_cuda
- test__int_mm_k_32_n_16_use_transpose_a_True_use_transpose_b_False_cuda
- test__int_mm_k_32_n_16_use_transpose_a_True_use_transpose_b_True_cuda
- test__int_mm_k_32_n_32_use_transpose_a_False_use_transpose_b_False_cuda
- test__int_mm_k_32_n_32_use_transpose_a_False_use_transpose_b_True_cuda
- test__int_mm_k_32_n_32_use_transpose_a_True_use_transpose_b_False_cuda
- test__int_mm_k_32_n_32_use_transpose_a_True_use_transpose_b_True_cuda

cc @jeffdaily @sunway513 @jithunnair-amd @pruthvistony @ROCmSupport @dllehr-amd @jataylo @hongxiayang

pytorch-bot · 2024-03-21T20:12:58Z

🔗 Helpful Links

🧪 See artifacts and rendered test results at hud.pytorch.org/pr/122431

📄 Preview Python docs built from this PR
📄 Preview C++ docs built from this PR
❓ Need help or want to give feedback on the CI? Visit the bot commands wiki or our office hours

Note: Links to docs will display an error until the docs builds have been completed.

✅ No Failures

As of commit 89ca666 with merge base a046606 ():
💚 Looks good so far! There are no failures yet. 💚

This comment was automatically generated by Dr. CI and updates every 15 minutes.

aten/src/ATen/cuda/CUDABlas.cpp

torch/utils/hipify/cuda_to_hip_mappings.py

soulitzer · 2024-04-01T17:31:44Z

looks okay to me, but will let others do final review

jithunnair-amd · 2024-04-04T21:24:04Z

@malfet Please review. The changes are ROCm-specific

pruthvistony · 2024-04-18T18:39:47Z

UTs are passing on CI - https://github.com/pytorch/pytorch/actions/runs/8439471329/job/23117907758

pruthvistony · 2024-04-18T18:40:58Z

@pytorchbot rebase

pytorchmergebot · 2024-04-18T18:42:34Z

@pytorchbot started a rebase job onto refs/remotes/origin/viable/strict. Check the current status here

pytorchmergebot · 2024-04-18T18:42:39Z

Successfully rebased int_mm onto refs/remotes/origin/viable/strict, please pull locally before adding more changes (for example, via git checkout int_mm && git pull --rebase)

malfet · 2024-04-23T20:59:15Z

@pytorchbot merge

pytorchmergebot · 2024-04-23T21:01:07Z

Merge started

Your change will be merged once all checks pass (ETA 0-4 Hours).

Learn more about merging in the wiki.

Questions? Feedback? Please reach out to the PyTorch DevX Team

Advanced Debugging

Check the merge workflow status
here

Update ROCm-triton to use the AMD backend from https://github.com/openai/triton Note: `test__int_mm` can be enabled after #122431 is landed Co-authored-by: Pruthvi Madugundu <pruthvigithub@gmail.com> Co-authored-by: Nikita Shulga <2453524+malfet@users.noreply.github.com> Pull Request resolved: #121801 Approved by: https://github.com/nmacchioni, https://github.com/malfet

Update ROCm-triton to use the AMD backend from https://github.com/openai/triton Note: `test__int_mm` can be enabled after pytorch#122431 is landed Co-authored-by: Pruthvi Madugundu <pruthvigithub@gmail.com> Co-authored-by: Nikita Shulga <2453524+malfet@users.noreply.github.com> Pull Request resolved: pytorch#121801 Approved by: https://github.com/nmacchioni, https://github.com/malfet

Update ROCm-triton to use the AMD backend from https://github.com/openai/triton Note: `test__int_mm` can be enabled after #122431 is landed Co-authored-by: Pruthvi Madugundu <pruthvigithub@gmail.com> Co-authored-by: Nikita Shulga <2453524+malfet@users.noreply.github.com> Pull Request resolved: #121801 Approved by: https://github.com/nmacchioni, https://github.com/malfet

Update ROCm-triton to use the AMD backend from https://github.com/openai/triton Note: `test__int_mm` can be enabled after pytorch#122431 is landed Co-authored-by: Pruthvi Madugundu <pruthvigithub@gmail.com> Co-authored-by: Nikita Shulga <2453524+malfet@users.noreply.github.com> Pull Request resolved: pytorch#121801 Approved by: https://github.com/nmacchioni, https://github.com/malfet

The PR - fixes int_mm() /int8_gemm() integration with hipblasLT backend (require ROCm 6.0). - enables/fixes the following tests on Rocm - test__int_mm_k_16_n_16_use_transpose_a_False_use_transpose_b_False_cuda - test__int_mm_k_16_n_16_use_transpose_a_False_use_transpose_b_True_cuda - test__int_mm_k_16_n_16_use_transpose_a_True_use_transpose_b_False_cuda - test__int_mm_k_16_n_16_use_transpose_a_True_use_transpose_b_True_cuda - test__int_mm_k_16_n_32_use_transpose_a_False_use_transpose_b_False_cuda - test__int_mm_k_16_n_32_use_transpose_a_False_use_transpose_b_True_cuda - test__int_mm_k_16_n_32_use_transpose_a_True_use_transpose_b_False_cuda - test__int_mm_k_16_n_32_use_transpose_a_True_use_transpose_b_True_cuda - test__int_mm_k_32_n_16_use_transpose_a_False_use_transpose_b_False_cuda - test__int_mm_k_32_n_16_use_transpose_a_False_use_transpose_b_True_cuda - test__int_mm_k_32_n_16_use_transpose_a_True_use_transpose_b_False_cuda - test__int_mm_k_32_n_16_use_transpose_a_True_use_transpose_b_True_cuda - test__int_mm_k_32_n_32_use_transpose_a_False_use_transpose_b_False_cuda - test__int_mm_k_32_n_32_use_transpose_a_False_use_transpose_b_True_cuda - test__int_mm_k_32_n_32_use_transpose_a_True_use_transpose_b_False_cuda - test__int_mm_k_32_n_32_use_transpose_a_True_use_transpose_b_True_cuda Pull Request resolved: pytorch#122431 Approved by: https://github.com/pruthvistony, https://github.com/jithunnair-amd, https://github.com/malfet, https://github.com/atalman

Update ROCm-triton to use the AMD backend from https://github.com/openai/triton Note: `test__int_mm` can be enabled after pytorch#122431 is landed Co-authored-by: Pruthvi Madugundu <pruthvigithub@gmail.com> Co-authored-by: Nikita Shulga <2453524+malfet@users.noreply.github.com> Pull Request resolved: pytorch#121801 Approved by: https://github.com/nmacchioni, https://github.com/malfet

Update ROCm-triton to use the AMD backend from https://github.com/openai/triton Note: `test__int_mm` can be enabled after #122431 is landed Co-authored-by: Pruthvi Madugundu <pruthvigithub@gmail.com> Co-authored-by: Nikita Shulga <2453524+malfet@users.noreply.github.com> Pull Request resolved: #121801 Approved by: https://github.com/nmacchioni, https://github.com/malfet

This pull request enables the int_mm_error tests for rocm 6.0+ . since #122431 landed Pull Request resolved: #124999 Approved by: https://github.com/jeffdaily, https://github.com/malfet

petrex requested review from IvanYashchuk, jeffdaily, jithunnair-amd, lezcano and nikitaved as code owners March 21, 2024 20:12

pytorch-bot bot added module: rocm AMD GPU support for Pytorch release notes: linalg_frontend release notes category labels Mar 21, 2024

pytorchbot added the open source label Mar 21, 2024

jeffdaily added the ciflow/rocm Trigger "default" config CI on ROCm label Mar 21, 2024

soulitzer reviewed Mar 22, 2024

View reviewed changes

aten/src/ATen/cuda/CUDABlas.cpp Outdated Show resolved Hide resolved

ezyang added the triaged This issue has been looked at a team member, and triaged and prioritized into an appropriate module label Mar 22, 2024

pruthvistony added the rocm This tag is for PRs from ROCm team label Mar 25, 2024

pruthvistony requested changes Mar 26, 2024

View reviewed changes

torch/utils/hipify/cuda_to_hip_mappings.py Outdated Show resolved Hide resolved

petrex requested review from pruthvistony and soulitzer March 26, 2024 19:54

jithunnair-amd approved these changes Mar 28, 2024

View reviewed changes

jithunnair-amd requested a review from malfet March 28, 2024 21:33

soulitzer removed their request for review April 1, 2024 17:31

jataylo mentioned this pull request Apr 18, 2024

[ROCm] Triton upstream AMD backend integration #121801

Closed

pruthvistony approved these changes Apr 18, 2024

View reviewed changes

pruthvistony added the rocm priority high priority ROCm PRs from performance or other aspects label Apr 18, 2024

petrex added 3 commits April 18, 2024 18:42

skip SM90 check

f5bed46

test_int_mm fix

f5ba209

enable int_mm_out for Rocm > 6.0

0201595

petrex added 5 commits April 18, 2024 18:42

enable test on rocm

9522272

lint

7ce1bb1

fix cublasLtMatmul() result_ptr

f669069

[lint]fix trailing spaces

3dad451

clean up cuda to hip mapping

89ca666

pytorchmergebot force-pushed the int_mm branch from 003e68c to 89ca666 Compare April 18, 2024 18:42

malfet approved these changes Apr 23, 2024

View reviewed changes

pytorch-bot bot added the ciflow/trunk Trigger trunk jobs on your pull request label Apr 23, 2024

atalman approved these changes Apr 23, 2024

View reviewed changes

pytorchmergebot added the merging label Apr 23, 2024

pytorchmergebot added the Merged label Apr 24, 2024

pytorchmergebot closed this in 2e7b8ff Apr 24, 2024

pytorchmergebot removed the merging label Apr 24, 2024

petrex mentioned this pull request Apr 26, 2024

[ROCm] Enable int_mm_error tests for rocm 6.0+ #124999

Closed

jithunnair-amd added a commit to ROCm/pytorch that referenced this pull request Oct 23, 2024

Unskip test__int_mm since PR pytorch#122431 is merged

5e41248

Provide feedback

Saved searches

Use saved searches to filter your results more quickly

Uh oh!

[ROCm] Fix Int_mm() Integration with hipblasLT #122431

[ROCm] Fix Int_mm() Integration with hipblasLT #122431

Uh oh!

petrex commented Mar 21, 2024 •

edited by pytorch-bot bot

Loading

Uh oh!

pytorch-bot bot commented Mar 21, 2024 •

edited

Loading

Uh oh!

Uh oh!

Uh oh!

soulitzer commented Apr 1, 2024

Uh oh!

jithunnair-amd commented Apr 4, 2024

Uh oh!

pruthvistony commented Apr 18, 2024

Uh oh!

pruthvistony commented Apr 18, 2024

Uh oh!

pytorchmergebot commented Apr 18, 2024

Uh oh!

pytorchmergebot commented Apr 18, 2024

Uh oh!

malfet commented Apr 23, 2024

Uh oh!

pytorchmergebot commented Apr 23, 2024

Uh oh!

Reviewers

Assignees

Labels

Projects

Milestone

Development

Uh oh!

10 participants

[ROCm] Fix Int_mm() Integration with hipblasLT #122431

[ROCm] Fix Int_mm() Integration with hipblasLT #122431

Uh oh!

Conversation

petrex commented Mar 21, 2024 • edited by pytorch-bot bot Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Uh oh!

pytorch-bot bot commented Mar 21, 2024 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

🔗 Helpful Links

🧪 See artifacts and rendered test results at hud.pytorch.org/pr/122431

✅ No Failures

Uh oh!

Uh oh!

Uh oh!

soulitzer commented Apr 1, 2024

Uh oh!

jithunnair-amd commented Apr 4, 2024

Uh oh!

pruthvistony commented Apr 18, 2024

Uh oh!

pruthvistony commented Apr 18, 2024

Uh oh!

pytorchmergebot commented Apr 18, 2024

Uh oh!

pytorchmergebot commented Apr 18, 2024

Uh oh!

malfet commented Apr 23, 2024

Uh oh!

pytorchmergebot commented Apr 23, 2024

Merge started

Uh oh!

Reviewers

Assignees

Labels

Projects

Milestone

Development

Uh oh!

10 participants

petrex commented Mar 21, 2024 •

edited by pytorch-bot bot

Loading

pytorch-bot bot commented Mar 21, 2024 •

edited

Loading