ROCm/composable_kernel
1
0
Fork 0
mirror of https://github.com/ROCm/composable_kernel.git synced 2026-05-14 18:17:44 +00:00
Code Issues Packages Projects Releases Wiki Activity
1,189 Commits 743 Branches 38 Tags
abefa068fb6e212fe3c122c19803a61baeb3cadf
Commit Graph

1189 Commits

This Branch
This Branch
All Branches
Author SHA1 Message Date
Rostyslav Geyyer
abefa068fb Refactor tolerances for correctness check in gemm op (#1188) 
* Refactor tolerances for correctness check

* Update tolerances

* Update host-side gemm

* Update reference gemm call

[ROCm/composable_kernel commit: 363feb482d]
 2024-03-08 12:05:05 -08:00
Lisa
2436ee6261 Update link (#1186) 
[ROCm/composable_kernel commit: 0e28de9766]
 2024-03-07 10:09:17 -08:00
yhuiYH
6d33f036ee Update CODEOWNERS to use documentation group (#1190) 
Also had to remove a name

[ROCm/composable_kernel commit: adb3615d1a]
 2024-03-07 10:08:37 -08:00
dependabot[bot]
a00747d147 Bump rocm-docs-core from 0.35.0 to 0.35.1 in /docs/sphinx (#1187) 
Bumps [rocm-docs-core](https://github.com/RadeonOpenCompute/rocm-docs-core) from 0.35.0 to 0.35.1.
- [Release notes](https://github.com/RadeonOpenCompute/rocm-docs-core/releases)
- [Changelog](https://github.com/ROCm/rocm-docs-core/blob/develop/CHANGELOG.md)
- [Commits](https://github.com/RadeonOpenCompute/rocm-docs-core/compare/v0.35.0...v0.35.1)

---
updated-dependencies:
- dependency-name: rocm-docs-core
  dependency-type: direct:production
  update-type: version-update:semver-patch
...

Signed-off-by: dependabot[bot] <support@github.com>
Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com>

[ROCm/composable_kernel commit: 1ddc8a841a]
 2024-03-05 21:55:01 -08:00
Paul Fultz II
871810b885 Add host lib (#1134) 
* Format

* Format

* Format

* Remove const

* Use the right template

* Format

* Format

* add row/col instances

* Add missing file

* fixed

* Format

* Updates

* Format

* fixed rrr layout

* Format

* Update test and embed modules

* Restore older version

* Update year

* Set -fPIC

* Format

* Use double for isnan

* rename host folder to codegen + minor fix

* add codegen CI test

* add option to build components without building CK

* fix the groovy syntax

* fix typo

* use the correct function for the codegen stage

---------

Co-authored-by: Jing Zhang <jizha@amd.com>
Co-authored-by: Illia Silin <98187287+illsilin@users.noreply.github.com>
Co-authored-by: illsilin <Illia.Silin@amd.com>

[ROCm/composable_kernel commit: 8eff4d62b6]
 2024-03-05 17:08:43 -08:00
Illia Silin
7241d5ad6a [CI] Add CI build and test stage on MI300. (#1185) 
[ROCm/composable_kernel commit: cf86621170]
 2024-03-05 10:42:16 -08:00
Rostyslav Geyyer
19366f13cd Fix example_gemm_xdl_fp8 (#1183) 
[ROCm/composable_kernel commit: 9ce18b045d]
 2024-03-01 16:42:15 -08:00
Rostyslav Geyyer
12b6143968 Update clipping for fp8/bf8 conversion (#1182) 
* Update clipping for fp8 conversion

* Add clipping for bf8 conversion

* Format

[ROCm/composable_kernel commit: acfb339238]
 2024-03-01 10:30:38 -08:00
amoskvic
4256edcdd2 Style improvement: improving type alias usage consistency in gemm-related client examples. Also copyright year update for all client examples. (#1180) 
Co-authored-by: Arseny Moskvichev <amoskvic@amd.com>

[ROCm/composable_kernel commit: a776978cbe]
 2024-02-28 16:39:03 -08:00
Illia Silin
048872f7c8 Clip fp8 to +/-240 on all targets. (#1172) 
* clip fp8 to +/-240 on all targets

* if inputs to fp8 conversion are +/-inf, they remain unaltered

* increase tolerance for test_elementwise_layernorm to prevent false errors

* change the input values for gemm examples to floats

* reduce gemm example float input values to prevent errors

* increase the tolerance for gemm examples

[ROCm/composable_kernel commit: d0c7b45150]
 2024-02-27 14:31:05 -06:00
Bartłomiej Kocot
6aa497d6c1 Remove unnecessary comments (#1177) 
[ROCm/composable_kernel commit: d909599729]
 2024-02-26 07:56:06 -06:00
dependabot[bot]
30a128ef84 Bump rocm-docs-core from 0.34.2 to 0.35.0 in /docs/sphinx (#1175) 
Bumps [rocm-docs-core](https://github.com/RadeonOpenCompute/rocm-docs-core) from 0.34.2 to 0.35.0.
- [Release notes](https://github.com/RadeonOpenCompute/rocm-docs-core/releases)
- [Changelog](https://github.com/ROCm/rocm-docs-core/blob/develop/CHANGELOG.md)
- [Commits](https://github.com/RadeonOpenCompute/rocm-docs-core/compare/v0.34.2...v0.35.0)

---
updated-dependencies:
- dependency-name: rocm-docs-core
  dependency-type: direct:production
  update-type: version-update:semver-minor
...

Signed-off-by: dependabot[bot] <support@github.com>
Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com>

[ROCm/composable_kernel commit: 2eb74a9c0c]
 2024-02-22 21:39:16 -08:00
jakpiase
b3a942c03c Add support for mixed precision bf16&int8 grouped gemm (#1166) 
* add support for mixed precision bf16&int8 grouped gemm

* fix gfx versions and add bf16 kbatch condition

* added reviewers comments

[ROCm/composable_kernel commit: 32d4be3d09]
 2024-02-21 10:35:35 +01:00
Bartłomiej Kocot
4d035536cb Extend permute scale support up to 6D (#1168) 
* Extend permute scale support up to 6D

* Fixes

* Fixes

* Update profiler/README.md

Co-authored-by: Lisa <lisajdelaney@gmail.com>

* Update profiler/README.md

Co-authored-by: Lisa <lisajdelaney@gmail.com>

* Update profiler/README.md

Co-authored-by: Lisa <lisajdelaney@gmail.com>

* Update profiler/README.md

Co-authored-by: Lisa <lisajdelaney@gmail.com>

* Update profiler/README.md

Co-authored-by: Lisa <lisajdelaney@gmail.com>

* Update profiler/README.md

Co-authored-by: Lisa <lisajdelaney@gmail.com>

* Update profiler/README.md

Co-authored-by: Lisa <lisajdelaney@gmail.com>

---------

Co-authored-by: Lisa <lisajdelaney@gmail.com>

[ROCm/composable_kernel commit: 66736edb95]
 2024-02-20 09:56:54 -08:00
dependabot[bot]
d35c8f19da Bump rocm-docs-core from 0.34.0 to 0.34.2 in /docs/sphinx (#1170) 
Bumps [rocm-docs-core](https://github.com/RadeonOpenCompute/rocm-docs-core) from 0.34.0 to 0.34.2.
- [Release notes](https://github.com/RadeonOpenCompute/rocm-docs-core/releases)
- [Changelog](https://github.com/ROCm/rocm-docs-core/blob/develop/CHANGELOG.md)
- [Commits](https://github.com/RadeonOpenCompute/rocm-docs-core/compare/v0.34.0...v0.34.2)

---
updated-dependencies:
- dependency-name: rocm-docs-core
  dependency-type: direct:production
  update-type: version-update:semver-patch
...

Signed-off-by: dependabot[bot] <support@github.com>
Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com>

[ROCm/composable_kernel commit: abac8b07dd]
 2024-02-16 07:48:52 -08:00
Illia Silin
8921adfb39 upgrade the ccache version and update links (#1169) 
[ROCm/composable_kernel commit: b9ab9f4b4f]
 2024-02-15 15:46:01 -08:00
Bartłomiej Kocot
22bfdc1168 Add optimized blockwise gemm using ck wrapper (#1157) 
* Add optimized blockwise gemm using ck wrapper

* Add basic gemm example

* Update docs

* Add tutorial for gemm using ck wrapper

* Add perf note

* edits

* Fix cmake

* Fixes

---------

Co-authored-by: Lisa Delaney <lisa.delaney@amd.com>

[ROCm/composable_kernel commit: 1e73adbc28]
 2024-02-13 17:04:36 +01:00
Bartłomiej Kocot
34bb3c7e31 Add bilinear conv fwd and bwd data instances (#1164) 
[ROCm/composable_kernel commit: bf98b47697]
 2024-02-13 11:49:05 +01:00
Illia Silin
baa25bcc4b add docker credentials before pushing image (#1165) 
[ROCm/composable_kernel commit: a78be3f69e]
 2024-02-12 16:11:32 -08:00
zjing14
8063a407e7 Optimizing fp8_fp16 mixedprec gemm (#1150) 
* add delayed cvt

* extend fp16 gemm_splitk instances for fp8_fp16 gemm

* add f8 example

* add 128 kperblk instances for fp8

* add kpb128 instance

* added more instances into kpb128

* clean code

* clean code

* fix

* fix

* fixed

* Update example/35_splitK_gemm/splitK_gemm_xdl_fp16_fp8.cpp

Co-authored-by: Bartłomiej Kocot <barkocot@amd.com>

* Update include/ck/tensor_operation/gpu/thread/threadwise_tensor_slice_transfer.hpp

Co-authored-by: Bartłomiej Kocot <barkocot@amd.com>

* Update library/src/tensor_operation_instance/gpu/gemm_splitk/device_gemm_xdl_splitk_f16_fp8_f16_mk_nk_mn_kpb128_instance.cpp

Co-authored-by: Bartłomiej Kocot <barkocot@amd.com>

---------

Co-authored-by: Jing Zhang <jizha@amd.com>
Co-authored-by: Bartłomiej Kocot <barkocot@amd.com>

[ROCm/composable_kernel commit: 602c4cc0d9]
 2024-02-12 09:45:42 -08:00
zjing14
cc25dc316f add generic instances for DeviceGemm_Xdl_CShuffle (#1161) 
* add generic instances

* clean code

[ROCm/composable_kernel commit: 94fbaac002]
 2024-02-09 10:20:53 -06:00
dependabot[bot]
ec7641b27b Bump rocm-docs-core from 0.33.2 to 0.34.0 in /docs/sphinx (#1163) 
Bumps [rocm-docs-core](https://github.com/RadeonOpenCompute/rocm-docs-core) from 0.33.2 to 0.34.0.
- [Release notes](https://github.com/RadeonOpenCompute/rocm-docs-core/releases)
- [Changelog](https://github.com/ROCm/rocm-docs-core/blob/develop/CHANGELOG.md)
- [Commits](https://github.com/RadeonOpenCompute/rocm-docs-core/compare/v0.33.2...v0.34.0)

---
updated-dependencies:
- dependency-name: rocm-docs-core
  dependency-type: direct:production
  update-type: version-update:semver-minor
...

Signed-off-by: dependabot[bot] <support@github.com>
Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com>

[ROCm/composable_kernel commit: 4581b5d504]
 2024-02-08 21:35:39 -08:00
Lakhinder Walia
77addd3131 fast_gelu: minor code reorg to enhance ref & gpu performance (#1162) 
[ROCm/composable_kernel commit: 1f306024d0]
 2024-02-07 19:24:51 -08:00
Illia Silin
0a90ec2237 Split-up instances to improve build times. (#1159) 
* split up splitk-gemm instances

* clean up some unused variables

* split the mk_kn_mn interwave splitk-gemm instances

* split up f16_f16_f16 mk_nk_mn splitk gemm instances

* fix clang format

* fix function names

* fix typo

* split up the 2 largest fp16*fp8 splitk gemm instances

* get rid of unused variables

* split up the largest splitk-gemm fp8*fp16 instance file

* split up the instances for xdl fp8 gemms

* split the headers for f16 and i8 for wmmma convolution instances

[ROCm/composable_kernel commit: 1b0fbaebbb]
 2024-02-07 12:47:12 -08:00
jakpiase
b41cf51951 Add support for mixed-precision f16bf16_int8 gemm (#1127) 
[ROCm/composable_kernel commit: ba86eadce5]
 2024-02-07 15:54:13 +01:00
dependabot[bot]
f59cd842c8 Bump rocm-docs-core from 0.33.1 to 0.33.2 in /docs/sphinx (#1160) 
Bumps [rocm-docs-core](https://github.com/RadeonOpenCompute/rocm-docs-core) from 0.33.1 to 0.33.2.
- [Release notes](https://github.com/RadeonOpenCompute/rocm-docs-core/releases)
- [Changelog](https://github.com/ROCm/rocm-docs-core/blob/develop/CHANGELOG.md)
- [Commits](https://github.com/RadeonOpenCompute/rocm-docs-core/compare/v0.33.1...v0.33.2)

---
updated-dependencies:
- dependency-name: rocm-docs-core
  dependency-type: direct:production
  update-type: version-update:semver-patch
...

Signed-off-by: dependabot[bot] <support@github.com>
Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com>

[ROCm/composable_kernel commit: 753cef783f]
 2024-02-06 21:24:32 -08:00
Bartlomiej Wroblewski
dd3c37d031 Implement direct loads split-K GEMM kernel (#1137) 
* WIP: Implement direct loads split-K GEMM kernel

* Clean the review

---------

Co-authored-by: Adam Osewski <19374865+aosewski@users.noreply.github.com>
Co-authored-by: Bartłomiej Kocot <barkocot@amd.com>

[ROCm/composable_kernel commit: 6951858221]
 2024-02-07 01:08:34 +01:00
dependabot[bot]
e000e4c442 Bump rocm-docs-core from 0.33.0 to 0.33.1 in /docs/sphinx (#1158) 
Bumps [rocm-docs-core](https://github.com/RadeonOpenCompute/rocm-docs-core) from 0.33.0 to 0.33.1.
- [Release notes](https://github.com/RadeonOpenCompute/rocm-docs-core/releases)
- [Changelog](https://github.com/ROCm/rocm-docs-core/blob/develop/CHANGELOG.md)
- [Commits](https://github.com/RadeonOpenCompute/rocm-docs-core/compare/v0.33.0...v0.33.1)

---
updated-dependencies:
- dependency-name: rocm-docs-core
  dependency-type: direct:production
  update-type: version-update:semver-patch
...

Signed-off-by: dependabot[bot] <support@github.com>
Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com>

[ROCm/composable_kernel commit: 6299621107]
 2024-02-06 09:24:34 -08:00
Illia Silin
6d7cc0a989 Delete any dangling images after building a new one. (#1155) 
* delete dangling docker images

* fix groovy syntax

* fix groovy syntax again

* try a different way to delete dangling images

[ROCm/composable_kernel commit: f0dd1da088]
 2024-02-05 10:34:47 -08:00
Illia Silin
b63148b2b1 Add support for more Navi2x and Navi3x models. (#1152) 
* add support for navi2x and navi3x models

* fix syntax

* use common macro for different mi300 architectures

[ROCm/composable_kernel commit: 180f16f9ac]
 2024-02-02 11:35:26 -08:00
Bartłomiej Kocot
1971822447 Extend gemm traits number for ck wrapper (#1153) 
[ROCm/composable_kernel commit: 171ca260b5]
 2024-02-02 11:25:54 -08:00
Illia Silin
6b17c34c28 add new performance tests for mixed fp16/fp8 gemms (#1151) 
[ROCm/composable_kernel commit: 112b691bb7]
 2024-01-31 13:27:17 -08:00
Bartłomiej Kocot
240330481c Add blockwise gemm to ck wrapper (#1139) 
* Add blockwise gemm to ck wrapper

* Add blockwise gemm traits

* Disable test_gemm for non xdl devices

* Fixes

* Add c layout descritpions

[ROCm/composable_kernel commit: f3b6c23ac5]
 2024-01-31 21:24:40 +01:00
Illia Silin
a2bc19ab69 update the name of the compiler staging branch (#1148) 
[ROCm/composable_kernel commit: 6651a124cc]
 2024-01-30 13:55:31 -08:00
Illia Silin
94cb9d6d85 turn off performance tests in CI by default until the infrastructure is fixed (#1147) 
[ROCm/composable_kernel commit: e7495e6bb7]
 2024-01-30 13:14:58 -08:00
dependabot[bot]
82dcf5a0a8 Bump rocm-docs-core from 0.31.0 to 0.33.0 in /docs/sphinx (#1144) 
Bumps [rocm-docs-core](https://github.com/RadeonOpenCompute/rocm-docs-core) from 0.31.0 to 0.33.0.
- [Release notes](https://github.com/RadeonOpenCompute/rocm-docs-core/releases)
- [Changelog](https://github.com/ROCm/rocm-docs-core/blob/develop/CHANGELOG.md)
- [Commits](https://github.com/RadeonOpenCompute/rocm-docs-core/compare/v0.31.0...v0.33.0)

---
updated-dependencies:
- dependency-name: rocm-docs-core
  dependency-type: direct:production
  update-type: version-update:semver-minor
...

Signed-off-by: dependabot[bot] <support@github.com>
Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com>

[ROCm/composable_kernel commit: 84832fc42d]
 2024-01-29 09:02:52 -08:00
Illia Silin
8d92b8d30d fix CK path for hipTensor (#1143) 
[ROCm/composable_kernel commit: 4a8297c28a]
 2024-01-25 17:05:43 -08:00
rocking
133d8e762f layernorm & groupnorm bwd gamma beta (#1133) 
* Add layernorm bwd gamma beta external api

* Add groupnorm external api

* Add layernorm bwd gamma beta profiler

* Add groupnorm bwd gamma beta ckProfiler

* Add layernorm & groupnorm bwd gamma beta test

* Fix groupnorm bwd gamma beta profiler bug

* Layernorm bwd weight client example

* Groupnorm bwd weight client example

* clang format

* Remove useless header

* Let inv_std be positive

* Rename to num_bytes and move this calculation outside the loop

[ROCm/composable_kernel commit: 28f68a5a99]
 2024-01-25 19:53:15 +08:00
Illia Silin
b5a3426d89 Fixing most of the cppcheck errors. (#1142) 
* fix cppcheck errors, first pass

* fix format

* fix returned value in examples

* add macro definitions for cppcheck

* fix the profile_gemm logic

* update the gemm profiler logic

* add more difinitions to cppcheck, fix couple more errors

* replace runtime error with message in device function

* fix a couple of int4 issues

* no return for fill function

* fix errors in data_types.hpp

* fix format

* fix few remaining errors

* fix errors in data_types.hpp

* fix last couple of errors in datat_types.hpp

[ROCm/composable_kernel commit: 180e572076]
 2024-01-24 13:47:48 -08:00
Bartłomiej Kocot
526905641c Fix possible linting errors in changelog (#1141) 
* Fix possible linting errors in changelog

* Update CHANGELOG.md

* Update CHANGELOG.md

* Update CHANGELOG.md

[ROCm/composable_kernel commit: 6169fbbdb3]
 2024-01-24 17:19:02 +01:00
zjing14
db32462b2d fixed return (#1138) 
[ROCm/composable_kernel commit: 1be4706366]
 2024-01-22 08:42:26 -08:00
Haocong WANG
ec7e5b1331 [GEMM] Optimization for MI200/300. (#1135) 
* Optimize GEMM on MI200/300:
1. Add new blockwise gemm pipeline
2. Add irregular splitk intances

* clang format + typo fix

* Fix a bug

[ROCm/composable_kernel commit: bb63b9732c]
 2024-01-19 07:02:22 -06:00
Bartłomiej Kocot
a69d7546b1 Add optimized copy to ck wrapper (#1126) 
* Add optimized copy to ck wrapper

* Example optimizations

* Fixes

* Move img2col test to client example

* Refactor example

* Fix docs

* Fixes

* Fix

* Fixes

* Fixes

* Fixes

* Fixes

* Fixes

---------

Co-authored-by: zjing14 <zhangjing14@gmail.com>

[ROCm/composable_kernel commit: 7e4eb4b800]
 2024-01-19 11:29:00 +01:00
Illia Silin
6d12ce4b13 add Adam to code owners (#1136) 
[ROCm/composable_kernel commit: 38882d8ab5]
 2024-01-18 19:20:40 -06:00
randyh62
ab91e6e73e Randyh docfix (#1130) 
* Update LICENSE

update to 2024

* Update index.rst

change license.md to license.html

* fix syntax

---------

Co-authored-by: illsilin <Illia.Silin@amd.com>

[ROCm/composable_kernel commit: 402a930a4a]
 2024-01-16 09:00:37 -08:00
Illia Silin
83d46ca80b add code owners (#1132) 
[ROCm/composable_kernel commit: c1b5b58192]
 2024-01-16 07:55:18 -08:00
Illia Silin
4612139045 Add cppcheck to CK CI. (#1125) 
* add cppcheck to the CK CI

* fix the path to CK source for cppcheck

* fix the path to CK source for cppcheck one more time

* fix the path to CK source for cppcheck third time

* change the path to ck_cppcheck.log

* install latest cppcheck from source

* fix bug in ck.hpp and use 20 threads for cppcheck

* create a switch to turn cppckeck on and off in CI

[ROCm/composable_kernel commit: e6d099c830]
 2024-01-15 09:11:45 -08:00
dependabot[bot]
f0da02cf0c Bump rocm-docs-core from 0.30.3 to 0.31.0 in /docs/sphinx (#1131) 
Bumps [rocm-docs-core](https://github.com/RadeonOpenCompute/rocm-docs-core) from 0.30.3 to 0.31.0.
- [Release notes](https://github.com/RadeonOpenCompute/rocm-docs-core/releases)
- [Changelog](https://github.com/RadeonOpenCompute/rocm-docs-core/blob/develop/CHANGELOG.md)
- [Commits](https://github.com/RadeonOpenCompute/rocm-docs-core/compare/v0.30.3...v0.31.0)

---
updated-dependencies:
- dependency-name: rocm-docs-core
  dependency-type: direct:production
  update-type: version-update:semver-minor
...

Signed-off-by: dependabot[bot] <support@github.com>
Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com>

[ROCm/composable_kernel commit: 636a31015a]
 2024-01-15 09:09:13 -07:00
dependabot[bot]
483dfa6f31 Bump sphinxcontrib-bibtex from 2.6.1 to 2.6.2 in /docs/sphinx (#1129) 
Bumps [sphinxcontrib-bibtex](https://github.com/mcmtroffaes/sphinxcontrib-bibtex) from 2.6.1 to 2.6.2.
- [Changelog](https://github.com/mcmtroffaes/sphinxcontrib-bibtex/blob/develop/CHANGELOG.rst)
- [Commits](https://github.com/mcmtroffaes/sphinxcontrib-bibtex/compare/2.6.1...2.6.2)

---
updated-dependencies:
- dependency-name: sphinxcontrib-bibtex
  dependency-type: direct:production
  update-type: version-update:semver-patch
...

Signed-off-by: dependabot[bot] <support@github.com>
Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com>

[ROCm/composable_kernel commit: 0ce417269d]
 2024-01-11 11:27:03 -07:00
Illia Silin
264de3901a Add an option to change the number of warm-up cycles and iterations. (#1124) 
* allow setting the number of warmup cycles and iterations for profiler

* fix the gemm_splitk and grouped_gemm examples

[ROCm/composable_kernel commit: 886d9eeb99]
 2024-01-09 09:43:08 -08:00