env_vars.rst 8.89 KB
Newer Older
1
2
3
4
5
6
Environment Variables
=====================

For parsing
---------------

Paul Fultz II's avatar
Paul Fultz II committed
7
.. envvar:: MIGRAPHX_TRACE_ONNX_PARSER
8
9
10
11
12

Set to "1", "enable", "enabled", "yes", or "true" to use.
Print debugging traces for the onnx parser.
Prints: initializers (if used), ONNX node operators, added MIGraphX instructions

Paul Fultz II's avatar
Paul Fultz II committed
13
.. envvar:: MIGRAPHX_DISABLE_FP16_INSTANCENORM_CONVERT
14
15
16
17
18
19
20
21
22

Set to "1", "enable", "enabled", "yes", or "true" to use.
Disables the conversion from fp16 to fp32 for the InstanceNormalization ONNX operator that MIGX does as a workaround for accuracy issues with reduce_mean/variance.
See ``parse_instancenorm.cpp`` for more details.


Matchers
------------

Paul Fultz II's avatar
Paul Fultz II committed
23
.. envvar:: MIGRAPHX_TRACE_MATCHES
24
25
26
27

Set to "1" to print the matcher that matches an instruction and the matched instruction.
Set to "2" and use the ``MIGRAPHX_TRACE_MATHCES_FOR`` flag to filter out results.

Paul Fultz II's avatar
Paul Fultz II committed
28
.. envvar:: MIGRAPHX_TRACE_MATCHES_FOR
29
30
31

Set to the name of any matcher and only traces for that matcher will be printed out.

Paul Fultz II's avatar
Paul Fultz II committed
32
.. envvar:: MIGRAPHX_VALIDATE_MATCHES
33
34
35
36
37
38
39

Set to "1", "enable", "enabled", "yes", or "true" to use.
Validate the module after finding the matches (runs ``module.validate()``).

Program Execution 
---------------------

Paul Fultz II's avatar
Paul Fultz II committed
40
.. envvar:: MIGRAPHX_TRACE_EVAL
41
42
43
44
45
46
47
48
49
50

Set to "1", "2", or "3" to use.
"1" prints the instruction run and the time taken.
"2" prints everything in "1" and a snippet of the output argument and some statistics (ex. min, max, mean) of the output.
"3" prints everything in "1" and the full output buffers.


Program Verification
------------------------

Paul Fultz II's avatar
Paul Fultz II committed
51
.. envvar:: MIGRAPHX_VERIFY_ENABLE_ALLCLOSE
52
53
54
55
56
57
58
59

Set to "1", "enable", "enabled", "yes", or "true" to use.
Uses ``allclose`` with the given ``atol`` and ``rtol`` for verifying ranges with ``driver verify`` or the tests that use ``migraphx/verify.hpp``.


Pass debugging or Pass controls
-----------------------------------

Paul Fultz II's avatar
Paul Fultz II committed
60
.. envvar:: MIGRAPHX_TRACE_ELIMINATE_CONTIGUOUS
61
62
63
64

Set to "1", "enable", "enabled", "yes", or "true" to use.
Debug print the instructions that have input ``contiguous`` instructions removed.

Paul Fultz II's avatar
Paul Fultz II committed
65
.. envvar:: MIGRAPHX_DISABLE_POINTWISE_FUSION
66
67
68
69

Set to "1", "enable", "enabled", "yes", or "true" to use.
Disables the ``fuse_pointwise`` compile pass.

Paul Fultz II's avatar
Paul Fultz II committed
70
.. envvar:: MIGRAPHX_DEBUG_MEMORY_COLORING
71
72
73
74

Set to "1", "enable", "enabled", "yes", or "true" to use.
Print debug statements for the ``memory_coloring`` pass.

Paul Fultz II's avatar
Paul Fultz II committed
75
.. envvar:: MIGRAPHX_TRACE_SCHEDULE
76
77
78
79

Set to "1", "enable", "enabled", "yes", or "true" to use.
Print debug statements for the ``schedule`` pass.

Paul Fultz II's avatar
Paul Fultz II committed
80
.. envvar:: MIGRAPHX_TRACE_PROPAGATE_CONSTANT
81
82
83
84

Set to "1", "enable", "enabled", "yes", or "true" to use.
Traces instructions replaced with a constant.

Paul Fultz II's avatar
Paul Fultz II committed
85
.. envvar:: MIGRAPHX_INT8_QUANTIZATION_PARAMS
86
87
88
89

Set to "1", "enable", "enabled", "yes", or "true" to use.
Print the quantization parameters in only the main module.

Paul Fultz II's avatar
Paul Fultz II committed
90
.. envvar:: MIGRAPHX_DISABLE_DNNL_POST_OPS_WORKAROUND
91
92
93
94

Set to "1", "enable", "enabled", "yes", or "true" to use.
Disable the DNNL post ops workaround.

Paul Fultz II's avatar
Paul Fultz II committed
95
.. envvar:: MIGRAPHX_DISABLE_MIOPEN_FUSION
96
97
98
99

Set to "1", "enable", "enabled", "yes", or "true" to use.
Disable MIOpen fusions.

Paul Fultz II's avatar
Paul Fultz II committed
100
.. envvar:: MIGRAPHX_DISABLE_SCHEDULE_PASS
101
102
103
104

Set to "1", "enable", "enabled", "yes", or "true" to use.
Disable the ``schedule`` pass.

Paul Fultz II's avatar
Paul Fultz II committed
105
.. envvar:: MIGRAPHX_DISABLE_REDUCE_FUSION
106
107
108
109

Set to "1", "enable", "enabled", "yes", or "true" to use.
Disable the ``fuse_reduce`` pass.

Paul Fultz II's avatar
Paul Fultz II committed
110
.. envvar:: MIGRAPHX_ENABLE_NHWC
111
112
113
114

Set to "1", "enable", "enabled", "yes", or "true" to use.
Enable the ``layout_nhwc`` pass.

Paul Fultz II's avatar
Paul Fultz II committed
115
.. envvar:: MIGRAPHX_ENABLE_CK
116
117
118
119
120

Set to "1", "enable", "enabled", "yes", or "true" to use.
Enable using the Composable Kernels library.
Should be used in conjunction with ``MIGRAPHX_DISABLE_MLIR=1``.

Paul Fultz II's avatar
Paul Fultz II committed
121
.. envvar:: MIGRAPHX_DISABLE_MLIR*
122
123
124
Set to "1", "enable", "enabled", "yes", or "true" to use.
Disable using the rocMLIR library.

Paul Fultz II's avatar
Paul Fultz II committed
125
.. envvar:: MIGRAPHX_ENABLE_EXTRA_MLIR
126
127
128
Set to "1", "enable", "enabled", "yes", or "true" to use.
Enables additional opportunities to use MLIR that may improve performance.

Paul Fultz II's avatar
Paul Fultz II committed
129
.. envvar:: MIGRAPHX_COPY_LITERALS
130
131
132
133
134
135
136

Set to "1", "enable", "enabled", "yes", or "true" to use.
Use ``hip_copy_to_gpu`` with a new ``literal`` instruction rather than use ``hip_copy_literal{}``.

Compilation traces
----------------------

Paul Fultz II's avatar
Paul Fultz II committed
137
.. envvar:: MIGRAPHX_TRACE_FINALIZE
138
139
140
141

Set to "1", "enable", "enabled", "yes", or "true" to use.
Debug print instructions during the ``module.finalize()`` step.

Paul Fultz II's avatar
Paul Fultz II committed
142
.. envvar:: MIGRAPHX_TRACE_COMPILE
143
144
145
146

Set to "1", "enable", "enabled", "yes", or "true" to use.
Print trace information for the graph compilation process.

Paul Fultz II's avatar
Paul Fultz II committed
147
.. envvar:: MIGRAPHX_TRACE_PASSES
148
149
150
151

Set to "1", "enable", "enabled", "yes", or "true" to use.
Print the compile pass and the program after the pass.

Paul Fultz II's avatar
Paul Fultz II committed
152
.. envvar:: MIGRAPHX_TIME_PASSES
153
154
155
156
157
158
159
160

Set to "1", "enable", "enabled", "yes", or "true" to use.
Time the compile passes.


GPU Kernels JIT compilation debugging (applicable for both hiprtc and hipclang)
-----------------------------------------

Paul Fultz II's avatar
Paul Fultz II committed
161
.. envvar:: MIGRAPHX_TRACE_CMD_EXECUTE
162
163
164
165

Set to "1", "enable", "enabled", "yes", or "true" to use.
Print commands executed by the MIGraphX ``process``.

Paul Fultz II's avatar
Paul Fultz II committed
166
.. envvar:: MIGRAPHX_TRACE_HIPRTC
167
168
169
170

Set to "1", "enable", "enabled", "yes", or "true" to use.
Print HIPRTC options and C++ file executed.

Paul Fultz II's avatar
Paul Fultz II committed
171
.. envvar:: MIGRAPHX_DEBUG_SAVE_TEMP_DIR
172
173
174
175

Set to "1", "enable", "enabled", "yes", or "true" to use.
Make it so the created temporary directories are not deleted.

Paul Fultz II's avatar
Paul Fultz II committed
176
.. envvar:: MIGRAPHX_GPU_DEBUG
177
178
179
180

Set to "1", "enable", "enabled", "yes", or "true" to use.
Internally, this adds the option ``-DMIGRAPHX_DEBUG`` when compiling GPU kernels. It enables assertions and capture of source locations for the errors. 

Paul Fultz II's avatar
Paul Fultz II committed
181
.. envvar:: MIGRAPHX_GPU_DEBUG_SYM
182
183
184
185

Set to "1", "enable", "enabled", "yes", or "true" to use.
Adds the option ``-g`` when compiling HIPRTC.

Paul Fultz II's avatar
Paul Fultz II committed
186
.. envvar:: MIGRAPHX_GPU_DUMP_SRC
187
188
189
190

Set to "1", "enable", "enabled", "yes", or "true" to use.
Dump the HIPRTC source files compiled.

Paul Fultz II's avatar
Paul Fultz II committed
191
.. envvar:: MIGRAPHX_GPU_DUMP_ASM
192
193
194
195

Set to "1", "enable", "enabled", "yes", or "true" to use.
Dump the hip-clang assembly.

Paul Fultz II's avatar
Paul Fultz II committed
196
.. envvar:: MIGRAPHX_GPU_OPTIMIZE
197
198
199
200

Set the optimization mode for GPU compile (``-O`` option).
Defaults to ``-O3``.

Paul Fultz II's avatar
Paul Fultz II committed
201
.. envvar:: MIGRAPHX_GPU_COMPILE_PARALLEL
202
203
204
205

Set to the number of threads to use.
Compile GPU code in parallel with the given number of threads.

Paul Fultz II's avatar
Paul Fultz II committed
206
.. envvar:: MIGRAPHX_TRACE_NARY
207
208
209
210

Set to "1", "enable", "enabled", "yes", or "true" to use.
Print the ``nary`` device functions used.

Paul Fultz II's avatar
Paul Fultz II committed
211
.. envvar:: MIGRAPHX_ENABLE_HIPRTC_WORKAROUNDS
212
213
214
215

Set to "1", "enable", "enabled", "yes", or "true" to use.
Enable HIPRTC workarounds for bugs in HIPRTC.

Paul Fultz II's avatar
Paul Fultz II committed
216
.. envvar:: MIGRAPHX_USE_FAST_SOFTMAX
217
218
219
220

Set to "1", "enable", "enabled", "yes", or "true" to use.
Use the fast softmax optimization.

Paul Fultz II's avatar
Paul Fultz II committed
221
.. envvar:: MIGRAPHX_ENABLE_NULL_STREAM
222
223
224
225

Set to "1", "enable", "enabled", "yes", or "true" to use.
Allow using null stream for miopen and hipStream.

Paul Fultz II's avatar
Paul Fultz II committed
226
.. envvar:: MIGRAPHX_NSTREAMS
227
228
229
230

Set to the number of streams to use.
Defaults to 1.

Paul Fultz II's avatar
Paul Fultz II committed
231
.. envvar:: MIGRAPHX_TRACE_BENCHMARKING
232
233
234
235
236
237
238

Set to "1" to print benchmarching trace.
Set to "2" to print benchmarching trace with more detail.

MLIR vars
-------------

Paul Fultz II's avatar
Paul Fultz II committed
239
.. envvar:: MIGRAPHX_TRACE_MLIR
240
241
242
243

Set to "1" to trace MLIR and print any failures.
Set to "2" to additionally print all MLIR operations.

Paul Fultz II's avatar
Paul Fultz II committed
244
.. envvar:: MIGRAPHX_MLIR_USE_SPECIFIC_OPS
245
246
247
248

Set to the name of the operations you want to always use MLIR regardless of GPU architecture.
Accepts a list of operators separated by commas (ex: "fused", "convolution", "dot").

Paul Fultz II's avatar
Paul Fultz II committed
249
.. envvar:: MIGRAPHX_MLIR_TUNING_DB
250
251
252

Set to the path of the MLIR tuning database to load.

Paul Fultz II's avatar
Paul Fultz II committed
253
.. envvar:: MIGRAPHX_MLIR_TUNING_CFG
254
255
256
257

Set to the path of the tuning configuration.
Appends to tuning cfg file that could be used with rocMLIR tuning scripts.

Paul Fultz II's avatar
Paul Fultz II committed
258
.. envvar:: MIGRAPHX_MLIR_TUNE_EXHAUSTIVE
259
260
261
262

Set to "1", "enable", "enabled", "yes", or "true" to use.
Do exhaustive tuning for MLIR.

263
264
265
266
.. envvar:: MIGRAPHX_MLIR_TUNE_LIMIT

Set to an integer greater than 1.
Limits the number of solutions that MLIR will use for tuning.
267
268
269
270

CK vars
-----------

Paul Fultz II's avatar
Paul Fultz II committed
271
.. envvar:: MIGRAPHX_LOG_CK_GEMM
272
273
274
275

Set to "1", "enable", "enabled", "yes", or "true" to use.
Print Composable Kernels GEMM traces.

Paul Fultz II's avatar
Paul Fultz II committed
276
.. envvar:: MIGRAPHX_CK_DEBUG
277
278
279
280

Set to "1", "enable", "enabled", "yes", or "true" to use.
Always add the ``-DMIGRAPHX_CK_CHECK=1`` for compiling Composable Kernels operators.

Paul Fultz II's avatar
Paul Fultz II committed
281
.. envvar:: MIGRAPHX_TUNE_CK
282
283
284
285
286
287
288

Set to "1", "enable", "enabled", "yes", or "true" to use.
Use tuning for Composable Kernels.

Testing 
------------

Paul Fultz II's avatar
Paul Fultz II committed
289
.. envvar:: MIGRAPHX_TRACE_TEST_COMPILE
290
291
292
293
294
295

Set to the target that you want to trace the compilation of (ex. "gpu", "cpu").
Prints the compile trace for the given target for the verify tests.
This flag shouldn't be used in conjunction with ``MIGRAPHX_TRACE_COMPILE``.
For the verify tests only use ``MIGRAPHX_TRACE_TEST_COMPILE``.

Paul Fultz II's avatar
Paul Fultz II committed
296
.. envvar:: MIGRAPHX_TRACE_TEST
297
298
299
300

Set to "1", "enable", "enabled", "yes", or "true" to use.
Prints the reference and target programs even if the verify passed successfully.

Paul Fultz II's avatar
Paul Fultz II committed
301
.. envvar:: MIGRAPHX_DUMP_TEST
302
303
304

Set to "1", "enable", "enabled", "yes", or "true" to use.
Dumps verify tests to ``.mxr`` files.