"megatron/legacy/data/dataset_utils.py" did not exist on "6495927e7ccf4a67720fd81fe5013f752a71f45c"
README.md 6.66 KB
Newer Older
1
2
# Composable Kernel

3
4
5
The Composable Kernel (CK) library provides a programming model for writing performance-critical
kernels for machine learning workloads across multiple architectures (GPUs, CPUs, etc.). The CK library
uses general purpose kernel languages, such as HIP C++.
Sam Wu's avatar
Sam Wu committed
6

7
CK uses two concepts to achieve performance portability and code maintainability:
8
9

* A tile-based programming model
10
11
* Algorithm complexity reduction for complex machine learning (ML) operators. This uses an innovative
   technique called *Tensor Coordinate Transformation*.
12

Sam Wu's avatar
Sam Wu committed
13
![ALT](/docs/data/ck_component.png "CK Components")
14

15
The current CK library is structured into four layers:
Sam Wu's avatar
Sam Wu committed
16

17
18
19
20
* Templated Tile Operators
* Templated Kernel and Invoker
* Instantiated Kernel and Invoker
* Client API
21

Sam Wu's avatar
Sam Wu committed
22
23
![ALT](/docs/data/ck_layer.png "CK Layers")

24
## General information
Sam Wu's avatar
Sam Wu committed
25

26
To build our documentation locally, use the following code:
Sam Wu's avatar
Sam Wu committed
27

28
``` bash
Sam Wu's avatar
Sam Wu committed
29
cd docs
Sam Wu's avatar
Sam Wu committed
30
pip3 install -r sphinx/requirements.txt
Sam Wu's avatar
Sam Wu committed
31
32
python3 -m sphinx -T -E -b html -d _build/doctrees -D language=en . _build/html
```
33

34
35
You can find a list of our developers and contributors on our [Contributors](/CONTRIBUTORS.md) page.
page.
36

37
38
```note
If you use CK, cite us as follows:
Sam Wu's avatar
Sam Wu committed
39

40
41
* [Realizing Tensor Operators Using Coordinate Transformations and Tile Based Programming](???):
  This paper will be available on arXiv soon.
42
* [CITATION.cff](/CITATION.cff)
43
```
44

45
CK is released under the **[MIT license](/LICENSE)**.
Sam Wu's avatar
Sam Wu committed
46

47
## Building CK
48

49
50
We recommend building CK inside Docker containers, which include all necessary packages. Pre-built
Docker images are available on [DockerHub](https://hub.docker.com/r/rocm/composable_kernel/tags).
51

52
1. To build a new Docker image, use the Dockerfile provided with the source code:
53

54
55
56
    ```bash
    DOCKER_BUILDKIT=1 docker build -t ck:latest -f Dockerfile .
    ```
Sam Wu's avatar
Sam Wu committed
57

58
2. Launch the Docker container:
59

60
61
62
63
64
65
66
67
68
69
    ```bash
    docker run                                     \
    -it                                            \
    --privileged                                   \
    --group-add sudo                               \
    -w /root/workspace                             \
    -v ${PATH_TO_LOCAL_WORKSPACE}:/root/workspace  \
    ck:latest                                      \
    /bin/bash
    ```
Sam Wu's avatar
Sam Wu committed
70

71
3. Clone CK source code from the GitHub repository and start the build:
Chao Liu's avatar
Chao Liu committed
72

73
74
75
76
77
78
    ```bash
    git clone https://github.com/ROCmSoftwarePlatform/composable_kernel.git && \
    cd composable_kernel && \
    mkdir build && \
    cd build
    ```
Sam Wu's avatar
Sam Wu committed
79

80
81
82
    You must set the `GPU_TARGETS` macro to specify the GPU target architecture(s) you want
    to run CK on. You can specify single or multiple architectures. If you specify multiple architectures,
    use a semicolon between each; for example, `gfx908;gfx90a;gfx940`.
83

84
85
86
87
88
89
90
91
    ```bash
    cmake                                                                                             \
    -D CMAKE_PREFIX_PATH=/opt/rocm                                                                    \
    -D CMAKE_CXX_COMPILER=/opt/rocm/bin/hipcc                                                         \
    -D CMAKE_BUILD_TYPE=Release                                                                       \
    -D GPU_TARGETS="gfx908;gfx90a"                                                                    \
    ..
    ```
Sam Wu's avatar
Sam Wu committed
92

93
94
    If you don't set `GPU_TARGETS` on the cmake command line, CK is built for all GPU targets
    supported by the current compiler (this may take a long time).
95

96
4. Build the entire CK library:
97

98
99
100
    ```bash
    make -j
    ```
101

102
5. Install CK:
103

104
105
106
    ```bash
    make -j install
    ```
107

108
## Optional post-install steps
109

110
* Build examples and tests:
111

112
113
114
    ```bash
    make -j examples tests
    ```
Sam Wu's avatar
Sam Wu committed
115

116
117
118
119
120
* Build and run all examples and tests:

    ```bash
    make -j check
    ```
121

122
    You can find instructions for running each individual example in [example](/example).
123

124
* Build ckProfiler:
125

126
127
128
129
130
131
132
133
134
135
136
137
138
139
140
141
142
143
144
145
146
147
148
149
150
151
152
153
154
155
156
157
    ```bash
    make -j ckProfiler
    ```

    You can find instructions for running ckProfiler in [profiler](/profiler).

Note the `-j` option for building with multiple threads in parallel. This speeds up the build significantly.
Depending on the number of CPU cores and the amount of RAM on your system, you may want to
limit the number of threads. For example, if you have a 128-core CPU and 64 Gb of RAM.

By default, `-j` launches one thread per CPU core, which can cause the build to run out of memory and
crash. In such cases, you can reduce the number of threads to 32 by using `-j32`.

Additional cmake flags can be used to significantly speed-up the build:

* `INSTANCES_ONLY` (default is OFF) must be set to ON in order to build only the instances and library
  while skipping all tests, examples, and profiler. This is useful in cases when you plan to use CK as a
  dependency and don't plan to run any examples or tests.

* `DTYPES` (default is not set) can be set to any subset of "fp64;fp32;fp16;fp8;bf16;int8" to build
  instances of select data types only. The main default data types are fp32 and fp16; you can safely skip
  other data types.

* `DL_KERNELS` (default is OFF) must be set to ON in order to build instances, such as `gemm_dl` or
  `batched_gemm_multi_d_dl`. These instances are useful on architectures like the NAVI2x, as most
  other platforms have faster instances, such as `xdl` or `wmma`, available.

## Using sccache for building

The default CK Docker images come with a pre-installed version of sccache, which supports clang
being used as hip-compiler (" -x hip"). Using sccache can help reduce the time to re-build code from
hours to 1-2 minutes. In order to invoke sccache, you need to run:
Sam Wu's avatar
Sam Wu committed
158

159
```bash
160
 sccache --start-server
161
```
JD's avatar
JD committed
162

163
then add the following flags to the cmake command line:
Sam Wu's avatar
Sam Wu committed
164

165
```bash
166
 -DCMAKE_CXX_COMPILER_LAUNCHER=sccache -DCMAKE_C_COMPILER_LAUNCHER=sccache
167
168
```

169
170
171
You may need to clean up the build folder and repeat the cmake and make steps in order to take
advantage of the sccache during subsequent builds.

172
## Using CK as pre-built kernel library
Sam Wu's avatar
Sam Wu committed
173

174
You can find instructions for using CK as a pre-built kernel library in [client_example](/client_example).
JD's avatar
JD committed
175

176
## Contributing to CK
177

178
179
When you contribute to CK, make sure you run `clang-format` on all changed files. We highly
recommend using git hooks that are managed by the `pre-commit` framework. To install hooks, run:
180
181
182
183
184

```bash
sudo script/install_precommit.sh
```

185
186
With this approach, `pre-commit` adds the appropriate hooks to your local repository and
automatically runs `clang-format` (and possibly additional checks) before any commit is created.
187
188
189
190
191
192
193

If you need to uninstall hooks from the repository, you can do so by running the following command:

```bash
script/uninstall_precommit.sh
```

194
195
If you need to temporarily disable pre-commit hooks, you can add the `--no-verify` option to the
`git commit` command.