mirror of https://github.com/ROCm/composable_kernel.git synced 2026-05-13 01:36:06 +00:00

Files

Thomas Ning 6f27bc9872 Ck tile gemm cshuffle & CK Tile GEMM restructure (#1535 )

* ake the cshuffle compilable

* modify Mhe reference on gpu and cpu. Correaccess of cshuffle

* fix the cpu reference code

* Complete the in tile shuffle logic

* restructure the kernel template input

* change the naming pattern of ck_tile gemm pipeline

* Re-format files using remod.py

* Solve the fmha conflict with gemm

* Comment Addressed from Carlus

---------

Co-authored-by: Po Yen, Chen <PoYen.Chen@amd.com>

2024-10-10 18:02:22 +08:00

algorithm

[CK_TILE] FA bwd kernels optimization (#1397 )

2024-08-16 13:40:10 -07:00

arch

[CK_TILE] wa prec, remove sgpr offset for inline asm (#1356 )

2024-07-08 11:09:55 -07:00

container

Ck tile gemm cshuffle & CK Tile GEMM restructure (#1535 )

2024-10-10 18:02:22 +08:00

numeric

[CK_TILE] float -> bf16 inline asm rtn (#1482 )

2024-08-30 15:38:09 +08:00

tensor

Enable scratch memory workaround on ROCm 6.2 (#1486 )

2024-08-30 10:40:00 +08:00

utility

[CK_TILE] Add PagedAttention kernels (#1387 )

2024-08-28 20:50:43 +08:00

config.hpp

[CK_TILE] float -> bf16 inline asm rtn (#1482 )

2024-08-30 15:38:09 +08:00

README.md

introducing ck_tile! (#1216 )

2024-04-15 19:27:12 -05:00

README.md

ck_tile/core

ck_tile/core contains every basic functions and structures to create a GPU kernel using ck_tile. User should only include ck_tile/core.hpp this single header to use all the functionality. Everything is under ck_tile namespace. The coding style under this folder should be similar to std (snake_case for structure/function, Camel for template types...)

algorithm/
    coordinate transform and some other reusable algorithm
arch/
    contains some basic device building block like mma, buffer addressing, etc...
container/
    contains basic container data structure, array/sequence/tuple/...
numeric/
    data type, and data type related math
tensor/
    tensor descriptors and tile level API
utility/
    other utility function for both host/device