注意
点击此处下载完整示例代码
Inductor CPU 后端调试与性能分析¶
创建日期：2023 年 7 月 1 日 | 最后更新：2025 年 1 月 8 日 | 最后验证：2024 年 11 月 5 日
作者: Xuan Liao, Haozhe Zhu, Jiong Gong, Weihan Wang
概述¶
PyTorch 2.0 引入了名为 torch.compile 的编译 API。这项新功能通过由默认 Inductor 后端驱动的图级别优化，提供了比 eager 模式执行显著的加速。
本教程旨在深入探讨 torch.compile 的细节，从而深入介绍 Inductor CPU 后端的调试和性能分析。
同时，你可能还会找到关于 torch.compile 的相关教程，涵盖基本用法、全面的故障排除以及 GPU 特定的知识，例如GPU 性能分析。
我们将从一个引发编译问题和精度问题的激励性示例开始调试，通过演示调试过程来精确定位问题。
通过启用日志记录和探索底层生成的代码，你可以学习如何一步步缩小故障范围，最终找出根本原因。
接下来，我们将讨论如何对编译后的代码进行性能分析，并通过与 eager 模式的性能比较，阐述 torch.compile 相较于其 eager 对应物为何能提供额外的性能提升。
调试¶
这是一个使用 Inductor 运行 torch.compile 并将其结果与 eager 模式进行比较的简单示例
import torch
def foo1(x1, x2):
a = torch.neg(x1)
b = torch.maximum(x2, a)
y = torch.cat([b], dim=0)
return y
x1 = torch.randint(256, (1, 8), dtype=torch.uint8)
x2 = torch.randint(256, (8390, 8), dtype=torch.uint8)
compiled_foo1 = torch.compile(foo1)
result = compiled_foo1(x1, x2)
cpp 代码生成中 neg 的正确实现如下
def neg1(x):
return f"decltype({x})(-{x})"
为了演示调试过程，我们稍后会将该函数修改为一个错误的实现。
获取更多日志信息¶
默认情况下，如果你运行这个简单示例，将不会提供调试信息。为了获取更多有用的调试和日志信息，我们通常会添加一个 TORCH_COMPILE_DEBUG 环境变量，如下所示
TORCH_COMPILE_DEBUG=1 python xx.py
这将在输出日志中打印更多调试信息，并转储代码生成过程中产生的中间 IR。你可以在日志中找到转储的文件路径，如下所示
torch._inductor.debug: [WARNING] model___20 debug trace: /tmp/torchinductor_root/rx/crxfi2ybd7yp5sbj2pnhw33wfhtdw7wumvrobyp5sjvdui5ktjc2.debug
在此目录下，保存了以下文件用于调试目的
文件 |
描述 |
|---|---|
|
可执行的 FX 图，经过分解，在模式匹配之前 |
|
转换后的 FX 图，在模式匹配之后 |
|
融合前的 Inductor IR |
|
融合后的 Inductor IR |
|
生成的 Python 图代码，包含 C++/Triton 内核 |
注意，fx_graph_runnable.py 和 output_code.py 都是可运行和可编辑的，以便于调试。以下是文件中的主要代码片段，我们将 C++ 生成的代码行与 FX 代码行进行关联。
fx_graph_runnable:
def forward1(self, arg0_1, arg1_1):
neg = torch.ops.aten.neg.default(arg0_1); arg0_1 = None
maximum = torch.ops.aten.maximum.default(arg1_1, neg); arg1_1 = neg = None
clone = torch.ops.aten.clone.default(maximum); maximum = None
return (clone,)
output_code 中的 C++ 内核
import torch
from torch._inductor.async_compile import AsyncCompile
async_compile = AsyncCompile()
cpp_fused_cat_maximum_neg_0 = async_compile.cpp('''
#include "/tmp/torchinductor_root/gv/cgv6n5aotqjo5w4vknjibhengeycuattfto532hkxpozszcgxr3x.h"
extern "C" void kernel(const unsigned char* in_ptr0,
const unsigned char* in_ptr1,
unsigned char* out_ptr0)
{
{
#pragma GCC ivdep
for(long i0=static_cast<long>(0L); i0<static_cast<long>(8390L); i0+=static_cast<long>(1L))
{
#pragma GCC ivdep
for(long i1=static_cast<long>(0L); i1<static_cast<long>(8L); i1+=static_cast<long>(1L))
{
auto tmp0 = in_ptr0[static_cast<long>(i1 + (8L*i0))];
auto tmp1 = in_ptr1[static_cast<long>(i1)];
// Corresponding FX code line: neg = torch.ops.aten.neg.default(arg0_1); arg0_1 = None
auto tmp2 = decltype(tmp1)(-tmp1);
// Corresponding FX code line: maximum = torch.ops.aten.maximum.default(arg1_1, neg); arg1_1 = neg = None
auto tmp3 = max_propagate_nan(tmp0, tmp2);
// Corresponding FX code line: clone = torch.ops.aten.clone.default(maximum); maximum = None
out_ptr0[static_cast<long>(i1 + (8L*i0))] = tmp3;
}
}
}
}''')
确定错误组件¶
当遇到错误或精度问题时，找到 bug 的直接解决方案是缩小问题范围。首先要做的就是确定错误发生的组件。幸运的是，这可以通过更改 torch.compile 的后端轻松实现。
代码 |
描述 |
|---|---|
|
启用 Dynamo |
|
启用 Dynamo + AOT Autograd |
|
启用 Dynamo + AOT Autograd + Inductor |
如果模型在后端设置为 eager 或 aot_eager 时能成功运行，而在使用 inductor 时失败，我们可以将故障范围缩小到 Inductor。
编译错误¶
正如我们所知，图级别优化的演进链如下
torch.neg (Python) -> torch.ops.aten.neg.default (within FX graph) -> ops.neg (within IR node) -> tmp2 = -tmp1 (within C++ kernel)
如果你遇到编译错误，这意味着在编译输出代码中的 C++ 内核时出现了问题。这类错误表明在将 IR 节点降低到输出代码时引入了 bug。编译错误的根本原因通常会显示在回溯日志中。
def neg2(x):
return f"-{x}"
例如，neg 函数被修改如下
torch._dynamo.exc.BackendCompilerFailed: backend='inductor' raised:
CppCompileError: C++ compile error
/tmp/torchinductor_root/xg/cxga5tk3b4lkwoxyigrtocjp5s7vc5cg2ikuscf6bk6pjqip2bhx.cpp: In function ‘void kernel(const unsigned char*, const unsigned char*, unsigned char*)’:
/tmp/torchinductor_root/xg/cxga5tk3b4lkwoxyigrtocjp5s7vc5cg2ikuscf6bk6pjqip2bhx.cpp:17:57: error: no matching function for call to ‘max_propagate_nan(unsigned char&, int&)’
17 | auto tmp3 = max_propagate_nan(tmp0, tmp2);
| ^
In file included from /tmp/torchinductor_root/xg/cxga5tk3b4lkwoxyigrtocjp5s7vc5cg2ikuscf6bk6pjqip2bhx.cpp:2:
/tmp/torchinductor_root/gv/cgv6n5aotqjo5w4vknjibhengeycuattfto532hkxpozszcgxr3x.h:27:17: note: candidate: ‘template<class scalar_t> scalar_t max_propagate_nan(scalar_t, scalar_t)’
27 | inline scalar_t max_propagate_nan(scalar_t a, scalar_t b) {
| ^~~~~~~~~~~~~~~~~
/tmp/torchinductor_root/gv/cgv6n5aotqjo5w4vknjibhengeycuattfto532hkxpozszcgxr3x.h:27:17: note: template argument deduction/substitution failed:
/tmp/torchinductor_root/xg/cxga5tk3b4lkwoxyigrtocjp5s7vc5cg2ikuscf6bk6pjqip2bhx.cpp:17:57: note: deduced conflicting types for parameter ‘scalar_t’ (‘unsigned char’ and ‘int’)
17 | auto tmp3 = max_propagate_nan(tmp0, tmp2);
| ^
日志提供了以下编译错误，原因相当清晰。
让我们也看看输出代码中的相应 C++ 内核和 IR 节点。
include "/tmp/torchinductor_root/gv/cgv6n5aotqjo5w4vknjibhengeycuattfto532hkxpozszcgxr3x.h"
extern "C" void kernel(const unsigned char* in_ptr0,
const unsigned char* in_ptr1,
unsigned char* out_ptr0)
{
{
#pragma GCC ivdep
for(long i0=static_cast<long>(0L); i0<static_cast<long>(8390L); i0+=static_cast<long>(1L))
{
#pragma GCC ivdep
for(long i1=static_cast<long>(0L); i1<static_cast<long>(8L); i1+=static_cast<long>(1L))
{
auto tmp0 = in_ptr0[static_cast<long>(i1 + (8L*i0))];
auto tmp1 = in_ptr1[static_cast<long>(i1)];
auto tmp2 = -tmp1;
auto tmp3 = max_propagate_nan(tmp0, tmp2);
out_ptr0[static_cast<long>(i1 + (8L*i0))] = tmp3;
}
}
}
}
C++ 内核
buf0: SchedulerNode(ComputedBuffer)
buf0.writes = [MemoryDep('buf0', c0, {c0: 67120})]
buf0.unmet_dependencies = []
buf0.met_dependencies =
[ MemoryDep('arg0_1', c1, {c0: 8390, c1: 8}),
MemoryDep('arg1_1', c0, {c0: 67120})]
buf0.users = [NodeUser(node=OUTPUT, can_inplace=False)]
buf0.group.device = cpu
buf0.group.iteration = ((8390, 8), ())
buf0.sizes = ([8390, 8], [])
class buf0_loop_body:
var_ranges = {z0: 8390, z1: 8}
index0 = 8*z0 + z1
index1 = z1
def body(self, ops):
get_index = self.get_index('index0')
load = ops.load('arg1_1', get_index)
get_index_1 = self.get_index('index1')
load_1 = ops.load('arg0_1', get_index_1)
neg = ops.neg(load_1)
maximum = ops.maximum(load, neg)
get_index_2 = self.get_index('index0')
store = ops.store('buf0', get_index_2, maximum, None)
return store
IR 节点
根据回溯日志，编译错误是由 max_propagate_nan 输入数据类型不一致引起的。通过检查 C++ 内核，我们知道在进行 - 操作后，tmp2 不再是 long 类型，因为 tmp0 是 long 类型。我们可以轻松地将 C++ 内核中的 - 和 max_propagate_nan 分别与 IR 节点中的 ops.neg 和 ops.maximum 相匹配。
现在我们成功找到根本原因：cpp 代码生成中 ops.neg 的实现，它在执行 neg 时悄然改变了数据类型。
精度调试¶
否则，如果模型运行出现其他错误或精度问题，你可以使用 PyTorch 调试工具 Minifier。
Minifier 的核心思想是持续移除图的节点和输入，直到找到出现问题的最小图。它通过 4 种策略帮助自动生成一个最小化的有问题图：截断后缀、增量调试、消除死代码和移除未使用的输入。
现在我们将借助 Minifer 展示精度问题的调试过程。精度问题指的是 eager 和 inductor 后端的输出不同的情况。
from torch._dynamo.utils import same
def foo2(x1, x2):
a = torch.neg(x1)
b = torch.maximum(x2, a)
y = torch.cat([b], dim=0)
return y
x1 = torch.randn((1, 8), dtype=torch.float32)
x2 = torch.randn((8390, 8), dtype=torch.float32)
expected_result = foo2(x1, x2)
compiled_foo2 = torch.compile(foo2)
actual_result = compiled_foo2(x1, x2)
assert same(expected_result, actual_result) == True
例如，我们像这样修改示例
def neg3(x):
return f"decltype({x})(2 * {x})"
并同时修改 neg 函数
torch._dynamo.utils: [ERROR] Accuracy failed: allclose not within tol=0.0001
Traceback (most recent call last):
File "test_script.py", line 18, in <module>
assert same(expected_result, actual_result) == True
AssertionError
将会引发一个精度问题，如下所示
TORCHDYNAMO_REPRO_AFTER="aot" TORCHDYNAMO_REPRO_LEVEL=4 python xx.py
要使用 Minifier 调试精度问题，需要两个环境变量
Started off with 6 nodes
Trying granularity 2
Strategy: Truncate suffix (G: 2) (6 nodes, 2 inputs)
SUCCESS: Went from 6 to 4 nodes
Trying granularity 4
Strategy: Remove unused inputs (G: 4) (4 nodes, 2 inputs)
SUCCESS: Went from 4 to 3 nodes
这会给我们提供日志信息，展示最小化的步骤
def forward2(self, arg0_1):
neg = torch.ops.aten.neg.default(arg0_1); arg0_1 = None
return (neg,)
运行后，我们得到最终最小化的图，其中包含目标节点 neg
有关 Minifier 的更多使用详情，请参阅故障排除。
性能分析¶
export KMP_BLOCKTIME=1
export KMP_SETTINGS=1
export KMP_AFFINITY=granularity=fine,compact,1,0
export LD_PRELOAD=${CONDA_PREFIX:-"$(dirname $(which conda))/../"}/lib/libiomp5.so:${CONDA_PREFIX:-"$(dirname $(which conda))/../"}/lib/libjemalloc.so
export MALLOC_CONF="oversize_threshold:1,background_thread:true,metadata_thp:auto,dirty_decay_ms:-1,muzzy_decay_ms:-1"
numactl -C 0-31 -m 0 python bench.py
# bench.py
from transformers import MobileBertForQuestionAnswering
# Initialize an eager model
model = MobileBertForQuestionAnswering.from_pretrained("csarron/mobilebert-uncased-squad-v2")
seq_length = 128
bs = 128
vocab_size = model.config.vocab_size
input = torch.randint(0, vocab_size, (bs, seq_length), dtype=torch.int64)
input_dict = {"input_ids": input}
# Initialize the inductor model
compiled_model = torch.compile(model)
with torch.no_grad():
compiled_model(**input_dict)
NUM_ITERS=50
import timeit
with torch.no_grad():
# warmup
for _ in range(10):
model(**input_dict)
eager_t = timeit.timeit("model(**input_dict)", number=NUM_ITERS, globals=globals())
with torch.no_grad():
# warmup
for _ in range(10):
compiled_model(**input_dict)
inductor_t = timeit.timeit("compiled_model(**input_dict)", number=NUM_ITERS, globals=globals())
# print(f"eager use: {eager_t * 1000 / NUM_ITERS} ms/iter")
# print(f"inductor use: {inductor_t * 1000 / NUM_ITERS} ms/iter")
# print(f"speed up ratio: {eager_t / inductor_t}")
在本节中，我们将演示如何对使用 Inductor CPU 后端编译的模型进行性能分析。在下面的示例中，我们使用 eager 模式和 Inductor 图模式对 Hugging Face Transformer 模型 MobileBertForQuestionAnswering 进行基准测试。基准测试后会打印 Inductor 的执行时间和加速比。我们使用 英特尔(R) Xeon(R) Platinum 8358 CPU @ 2.60GHz 并在第一个 socket 上运行基准测试，以演示本节中的优化。我们将以下环境变量设置为在 英特尔(R) CPU 上进行基准测试的最佳实践。
eager use: 802.1023553796113 ms/iter
inductor use: 339.95180135127157 ms/iter
speed up ratio: 2.359459053287382
输出
在我们自己的测试中，我们发现 Inductor CPU 后端将模型加速了约 2.355 倍。
from torch._inductor import config
config.cpp.enable_kernel_profile = True
接下来，让我们深入研究操作层面的性能，以了解加速来自何处。PyTorch Profiler 是一个很好的工具来帮助我们。Inductor CPU 后端支持使用 enable_kernel_profile 配置选项向 profiler 报告融合内核的时间
# bench.py
from torch.profiler import profile, schedule, ProfilerActivity
RESULT_DIR = "./prof_trace"
my_schedule = schedule(
skip_first=10,
wait=5,
warmup=5,
active=1,
repeat=5)
def trace_handler(p):
output = p.key_averages().table(sort_by="self_cpu_time_total", row_limit=20)
# print(output)
p.export_chrome_trace(f"{RESULT_DIR}/{p.step_num}.json")
for _ in range(10):
model(**input_dict) # compiled_model(**input_dict) to get inductor model profiling
total = 0
with profile(
activities=[ProfilerActivity.CPU],
schedule=my_schedule,
on_trace_ready=trace_handler
) as p:
for _ in range(50):
model(**input_dict) # compiled_model(**input_dict) to get inductor model profiling
p.step()
按照PyTorch Profiler 中的步骤，我们可以获得性能分析表和跟踪文件。
------------------------- ------------ ------------ ------------
Name CPU total % CPU total # of Calls
------------------------- ------------ ------------ ------------
aten::addmm 45.73% 370.814ms 362
aten::add 19.89% 161.276ms 363
aten::copy_ 14.97% 121.416ms 488
aten::mul 9.02% 73.154ms 194
aten::clamp_min 8.81% 71.444ms 96
aten::bmm 5.46% 44.258ms 48
ProfilerStep* 100.00% 810.920ms 1
aten::div 2.89% 23.447ms 24
aten::_softmax 1.00% 8.087ms 24
aten::linear 46.48% 376.888ms 362
aten::clone 2.77% 22.430ms 98
aten::t 0.31% 2.502ms 362
aten::view 0.14% 1.161ms 850
aten::transpose 0.17% 1.377ms 386
aten::index_select 0.12% 952.000us 3
aten::expand 0.12% 986.000us 458
aten::matmul 8.31% 67.420ms 48
aten::cat 0.09% 703.000us 1
aten::as_strided 0.08% 656.000us 963
aten::relu 8.86% 71.864ms 96
------------------------- ------------ ------------ ------------
Self CPU time total: 810.920ms
我们得到了 eager 模式模型的以下性能分析表（省略了一些列）
----------------------------------------------- ------------ ------------ ------------
Name CPU total % CPU total # of Calls
----------------------------------------------- ------------ ------------ ------------
mkl::_mkl_linear 68.79% 231.573ms 362
aten::bmm 8.02% 26.992ms 48
ProfilerStep* 100.00% 336.642ms 1
graph_0_cpp_fused_constant_pad_nd_embedding_0 0.27% 915.000us 1
aten::empty 0.27% 911.000us 362
graph_0_cpp_fused__mkl_linear_add_mul_relu_151 0.27% 901.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_226 0.27% 899.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_361 0.27% 898.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_121 0.27% 895.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_31 0.27% 893.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_76 0.26% 892.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_256 0.26% 892.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_346 0.26% 892.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_241 0.26% 891.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_316 0.26% 891.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_91 0.26% 890.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_106 0.26% 890.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_211 0.26% 890.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_61 0.26% 889.000us 1
graph_0_cpp_fused__mkl_linear_add_mul_relu_286 0.26% 889.000us 1
----------------------------------------------- ------------ ------------ ------------
Self CPU time total: 336.642ms
类似地，我们也获得了使用 Inductor 编译的模型的性能分析表（省略了一些列）
从 eager 模型的性能分析表中，我们可以看到最耗时的操作是 [aten::addmm, aten::add, aten::copy_, aten::mul, aten::clamp_min, aten::bmm]。与 inductor 模型的性能分析表相比，我们注意到一个 mkl::_mkl_linear 条目和多个形式为 graph_0_cpp_fused_* 的融合内核。它们是 inductor 模型进行的主要优化。让我们分别讨论它们。(1) 关于 mkl::_mkl_linear：你可能会注意到对该内核的调用次数是 362 次，这与 eager 模型性能分析表中的 aten::linear 完全相同。aten::linear 的 CPU 总时间为 376.888ms，而 mkl::_mkl_linear 的时间为 231.573ms。这表明“linear”部分的加速约为 1.63 倍。加速主要来自将权重张量打包成块状内存格式，并在 Inductor CPU 后端中调用cblas_sgemm_compute，以在 GEMM 计算期间获得更好的缓存行为。
cpp_fused__mkl_linear_add_mul_relu_151 = async_compile.cpp('''
#include <ATen/record_function.h>
#include "/tmp/torchinductor_root/lr/clrlgu27q4ggd472umdzwsu6qcpqxcuusjxqvx2hwitjbujiiz7z.h"
extern "C" void kernel(float* in_out_ptr0,
const float* in_ptr0,
const float* in_ptr1,
const float* in_ptr2,
const float* in_ptr3)
{
RECORD_FUNCTION("graph_0_cpp_fused__mkl_linear_add_mul_relu_151", c10::ArrayRef<c10::IValue>({}));
#pragma omp parallel num_threads(32)
{
{
#pragma omp for
for(long i0=static_cast<long>(0L); i0<static_cast<long>(16384L); i0+=static_cast<long>(1L))
{
for(long i1=static_cast<long>(0L); i1<static_cast<long>(512L); i1+=static_cast<long>(8L))
{
auto tmp0 = at::vec::Vectorized<float>::loadu(in_ptr0 + static_cast<long>(i1 + (512L*i0)));
auto tmp1 = at::vec::Vectorized<float>::loadu(in_ptr1 + static_cast<long>(i1));
auto tmp3 = at::vec::Vectorized<float>::loadu(in_out_ptr0 + static_cast<long>(i1 + (512L*i0)));
auto tmp5 = at::vec::Vectorized<float>::loadu(in_ptr2 + static_cast<long>(i1));
auto tmp7 = at::vec::Vectorized<float>::loadu(in_ptr3 + static_cast<long>(i1));
auto tmp2 = tmp0 + tmp1;
auto tmp4 = tmp2 + tmp3;
auto tmp6 = tmp4 * tmp5;
auto tmp8 = tmp6 + tmp7;
tmp8.store(in_out_ptr0 + static_cast<long>(i1 + (512L*i0)));
}
}
}
}
}''')
(2) 关于其他内存密集型操作：在我们自己的测试中，eager/inductor 模型的端到端延迟分别为 802ms/339ms。因此，我们可以大致推断其他内存密集型操作的加速约为 3.94 倍。让我们阅读生成的代码，了解 inductor 是如何实现这一令人印象深刻的优化的。你可以在 output_code.py 中搜索 cpp_fused__mkl_linear_add_mul_relu_151 来找到生成的代码。
# bench.py
def func(arg_0, arg_1, arg_2, arg_3, arg_4):
add_0 = arg_0 + arg_1
add_1 = add_0 + arg_2
mul_1 = add_1 * arg_3
add_2 = mul_1 + arg_4
arg_2 = add_2
return arg_2
arg_0 = torch.rand(16384, 512)
arg_1 = torch.rand(1, 512)
arg_2 = torch.zeros(16384, 512)
arg_3 = torch.rand(1, 512)
arg_4 = torch.rand(1, 512)
input = (arg_0, arg_1, arg_2, arg_3, arg_4)
inductor_func = torch.compile(func)
with torch.no_grad():
inductor_func(*input)
import timeit
NUM_ITERS=100
with torch.no_grad():
# warmup
for _ in range(10):
func(*input)
eager_t = timeit.timeit("func(*input)", number=NUM_ITERS, globals=globals())
with torch.no_grad():
# warmup
for _ in range(10):
inductor_func(*input)
inductor_t = timeit.timeit("inductor_func(*input)", number=NUM_ITERS, globals=globals())
# print(f"eager use: {eager_t * 1000 / NUM_ITERS} ms/iter")
# print(f"inductor use: {inductor_t * 1000 / NUM_ITERS} ms/iter")
# print(f"speed up ratio: {eager_t / inductor_t}")
在本节中，我们将演示如何对使用 Inductor CPU 后端编译的模型进行性能分析。在下面的示例中，我们使用 eager 模式和 Inductor 图模式对 Hugging Face Transformer 模型 MobileBertForQuestionAnswering 进行基准测试。基准测试后会打印 Inductor 的执行时间和加速比。我们使用 英特尔(R) Xeon(R) Platinum 8358 CPU @ 2.60GHz 并在第一个 socket 上运行基准测试，以演示本节中的优化。我们将以下环境变量设置为在 英特尔(R) CPU 上进行基准测试的最佳实践。
eager use: 5.780875144992024 ms/iter
inductor use: 0.9588955780491233 ms/iter
speed up ratio: 6.0286805751604735
从上面的生成代码中，我们可以看到该内核对 [add, add, mul, add] 执行了典型的循环融合。这是一个内存限制的瓶颈，阻碍了良好的性能。为了更直观地了解这种优化，我们可以推断输入的大小和步长，并进一步对 [add, add, mul, add] 模式进行基准测试。
这只是一个示例。性能分析表显示，在该模型中，Inductor 自动融合了所有逐元素操作。你可以在 output_code.py 中阅读更多内核。
结论¶
本文档提供了关于 Inductor CPU 后端的深入教程。
通过激励性示例，我们逐步介绍了调试和性能分析的过程。核心思想是缩小问题范围。
我们逐步演示了如何深入探究问题并找到故障的根本原因，这得益于调试日志和 Minifier 工具的帮助。首先确定故障发生在哪个组件中，然后尝试生成能够重现故障的最小代码片段。
当 Inductor 的性能优于 eager 模式时，我们提供了一种可靠的性能分析方法。我们展示了如何使用 PyTorch Profiler 找到耗时的热点，并确定操作级别或内核级别的原因来解释这种现象。
