benchmark
benchmark copied to clipboard
PaddlePaddle
Optimize inference performance of ERNIE on CPU
负责人
@tensor-tang @GaoWei8
机器型号
6148
commit号
based on https://github.com/PaddlePaddle/benchmark/pull/164
初始性能
10个sample
I0820 10:33:36.597270 35686 inference.cc:211] Load 10 samples from /home/tangjian/ernie/Inference/c++/ernie/seq128_data/test_ds_10
I0820 10:33:37.552497 35686 inference.cc:351] Run 10 samples, average latency: 95.519 ms per sample.
I0820 10:33:37.552565 35686 inference.cc:356] Run 9 samples, average latency [exclude 1 warmup steps]: 89.8265 ms per sample.
profile 结果
Event Calls Total Min. Max. Ave. Ratio.
thread0::fc 740 625.813 0.013066 2.59008 0.845693 0.511826
thread0::load 202 269.789 0.009506 168.902 1.33559 0.220649
thread0::elementwise_add 380 78.8811 0.045364 29.1685 0.207582 0.0645135
thread0::transpose2 480 63.5249 0.089001 4.08297 0.132343 0.0519543
thread0::dropout 380 51.7229 0.01217 0.262424 0.136113 0.0423019
thread0::layer_norm 250 43.0832 0.150904 0.226994 0.172333 0.0352359
thread0::matmul 250 36.4239 0.033627 14.6704 0.145696 0.0297895
thread0::relu 120 22.7715 0.130891 1.77192 0.189762 0.0186238
thread0::scale 140 11.0508 0.006102 0.105016 0.0789342 0.00903797
thread0::softmax 120 9.73205 0.050275 0.451451 0.0811004 0.00795943
thread0::reshape2 480 4.47205 0.006964 0.022523 0.00931677 0.0036575
thread0::lookup_table 30 2.67894 0.074823 0.105928 0.089298 0.00219099
thread0::stack 10 1.43889 0.130984 0.154692 0.143889 0.00117681
thread0::tanh 10 0.986778 0.084346 0.191761 0.0986778 0.000807043
thread0::slice 10 0.12234 0.009367 0.033865 0.012234 0.000100057
thread0::feed 40 0.109835 0.001013 0.005219 0.00274588 8.98293e-05
thread0::fetch 10 0.106458 0.006874 0.011848 0.0106458 8.70674e-05
TODO
- [ ] ~~去掉load @tensor-tang，不会统计进预测时间，可以忽略~~
- [ ] dropout多线程 @gaowei8
- [ ] fuse @intel
全量结果
I0821 06:53:24.524538 37140 inference.cc:351] Run 5010 samples, average latency: 88.4711 ms per sample.
I0821 06:53:24.524575 37140 inference.cc:356] Run 5009 samples, average latency [exclude 1 warmup steps]: 88.465 ms per sample.
W0821 06:53:25.474104 37140 profiler.cc:89] CUDA CUPTI is not enabled
-------------------------> Profiling Report <-------------------------
Place: CPU
Time unit: ms
Sorted by total time in descending order in the same thread
Event Calls Total Min. Max. Ave. Ratio.
thread0::fc 370740 272578 0.011775 19.7095 0.735228 0.616268
thread0::elementwise_add 190380 48865.5 0.114381 1.73744 0.256674 0.110479
thread0::transpose2 240480 32294.4 0.077649 0.449438 0.134291 0.0730139
thread0::relu 60120 21220.9 0.263376 1.20742 0.352976 0.047978
thread0::dropout 190380 18181.7 0.009459 0.272083 0.0955024 0.0411068
thread0::layer_norm 125250 16305.7 0.106911 0.322171 0.130185 0.0368654
thread0::softmax 60120 13706.9 0.214626 0.525724 0.227993 0.0309898
thread0::matmul 125250 9322.28 0.022049 3.70151 0.0744294 0.0210766
thread0::scale 70140 4563.8 0.006707 0.204413 0.065067 0.0103182
thread0::reshape2 240480 2619.45 0.007166 0.078659 0.0108926 0.00592229
thread0::lookup_table 15030 1331 0.070208 0.277766 0.0885562 0.00300924
thread0::stack 5010 684.089 0.046387 0.335377 0.136545 0.00154665
thread0::load 202 268.195 0.010278 152.841 1.3277 0.000606358
thread0::tanh 5010 184.327 0.029071 0.183187 0.0367917 0.000416741
thread0::fetch 5010 63.0403 0.009063 0.024473 0.0125829 0.000142527
thread0::feed 20040 58.3227 0.001681 0.020759 0.00291031 0.000131861
thread0::slice 5010 56.4965 0.009148 0.039881 0.0112767 0.000127732
单线程 10个数据
Run 10 samples, average latency: 352.382 ms per sample.
I0821 12:07:01.099709 319 inference.cc:359] Run 9 samples, average latency [exclude 1 warmup steps]: 355.139 ms per sample.
W0821 12:07:01.102002 319 profiler.cc:89] CUDA CUPTI is not enabled
-------------------------> Profiling Report <-------------------------
Place: CPU
Time unit: ms
Sorted by total time in descending order in the same thread
Event Calls Total Min. Max. Ave. Ratio.
thread0::fc 740 3074.14 0.01086 20.773 4.15425 0.807906
thread0::load 202 286.345 0.010896 166.914 1.41755 0.0752534
thread0::elementwise_add 380 115.306 0.116436 1.67314 0.303438 0.0303033
thread0::matmul 250 83.0643 0.029765 3.11798 0.332257 0.0218299
thread0::transpose2 480 65.5743 0.075499 0.488148 0.136613 0.0172334
thread0::dropout 380 45.1451 0.01092 0.46845 0.118803 0.0118645
thread0::relu 120 43.6511 0.262822 1.60205 0.363759 0.0114718
thread0::layer_norm 250 33.2642 0.110461 0.214669 0.133057 0.00874208
thread0::softmax 120 29.9144 0.215222 0.439892 0.249287 0.00786172
thread0::scale 140 15.0192 0.005788 0.402538 0.10728 0.00394716
thread0::reshape2 480 5.56671 0.007877 0.046785 0.0115973 0.00146297
thread0::lookup_table 30 5.1676 0.070664 0.60177 0.172253 0.00135808
thread0::stack 10 1.93474 0.128399 0.227941 0.193474 0.000508463
thread0::tanh 10 0.542741 0.035165 0.146138 0.0542741 0.000142636
thread0::slice 10 0.160324 0.011435 0.042496 0.0160324 4.21343e-05
thread0::feed 40 0.147486 0.001027 0.01444 0.00368715 3.87604e-05
thread0::fetch 10 0.126759 0.007186 0.018538 0.0126759 3.33132e-05
20线程，10个sample
I0821 12:07:48.280498 335 inference.cc:354] Run 10 samples, average latency: 64.4593 ms per sample.
I0821 12:07:48.280521 335 inference.cc:359] Run 9 samples, average latency [exclude 1 warmup steps]: 62.4416 ms per sample.
W0821 12:07:48.282289 335 profiler.cc:89] CUDA CUPTI is not enabled
-------------------------> Profiling Report <-------------------------
Place: CPU
Time unit: ms
Sorted by total time in descending order in the same thread
Event Calls Total Min. Max. Ave. Ratio.
thread0::fc 740 432.367 0.012076 1.36456 0.58428 0.470114
thread0::load 202 277.494 0.011123 168.398 1.37373 0.30172
thread0::dropout 380 51.7348 0.011015 0.214377 0.136144 0.0562514
thread0::layer_norm 250 40.4131 0.153628 0.20805 0.161652 0.0439413
thread0::elementwise_add 380 32.8085 0.044845 3.63877 0.0863381 0.0356728
thread0::matmul 250 27.6031 0.030286 9.82356 0.110413 0.030013
thread0::transpose2 480 22.741 0.03887 0.200149 0.0473772 0.0247264
thread0::scale 140 10.2398 0.005754 0.100433 0.0731412 0.0111337
thread0::relu 120 8.8092 0.060546 0.175747 0.07341 0.00957827
thread0::softmax 120 6.71831 0.042472 0.268144 0.0559859 0.00730484
thread0::reshape2 480 4.31378 0.007102 0.021215 0.00898705 0.00469039
thread0::lookup_table 30 2.57996 0.07498 0.093309 0.0859988 0.0028052
thread0::stack 10 1.11586 0.104187 0.1304 0.111586 0.00121328
thread0::tanh 10 0.42612 0.030598 0.114753 0.042612 0.000463322
thread0::slice 10 0.130643 0.009965 0.03432 0.0130643 0.000142049
thread0::feed 40 0.106352 0.001016 0.004525 0.0026588 0.000115637
thread0::fetch 10 0.104856 0.006945 0.011369 0.0104856 0.00011401
baseline： 在CPU上面的测试数据 @徐屹， 82.7681ms intel BRTR 28.62ms
优化迭代：
82.7681 ms → 60.3766 ms (提升27%)
ERNIE BERT 分析.pdf @bingyanghuang 的分析。
主要结论为
- 单线程
- 状态
- 去掉冗余的reshape和transpose后，在BERT base上的性能提升1.8%
- 在FP32 上进一步优化空间较少
- 下一步
- 使用INT8量化进一步优化FC，以达到较为大幅度的性能提升
- 状态
- 多线程
- 状态
- 与竞品的2倍左右的差距初步怀疑是paddle多线程中存在一些bug导致CPU空转时间较长
- 下一步
- 查看Vtune log找到paddle 多线程缓慢的root cause
- 总结竞品所使用的优化方式
- 状态
一个需求：
layernorm 多线程
Top Serial Hotspots (outside parallel regions)
Function Module Serial CPU Time
-------------------------------------------------- -------------- ---------------
paddle::operators::jit::more::intrinsic::LayerNorm inference 1.338s
__kmp_join_call libiomp5.so 0.782s
__GI_ libc.so.6 0.413s
_mm256_storeu_ps inference 0.296s
std::string::string libstdc++.so.6 0.290s
[Others] N/A 4.855s
Next step
1. mkl sgemm 尺寸测试
在ernie中
在paddle中单独测试，包含了每次iteration刷新矩阵X。
2. 去掉锁。https://github.com/PaddlePaddle/Paddle/pull/19624
3. 更优可以layernorm的多线程问题。
@tensor-tang Ernie关于多线程性能差的问题，我们这边做了如下测试：
- 环境变量 disable HT and Turbo Boost
export OMP_NUM_THREADS=20
export KMP_BLOCKTIME=1
export KMP_AFFINITY=granularity=fine,compact,1,0
numactl --cpunodebind=0 --membind=0 CMD
- Intel-MKLML 小包 UT
| GEMM Size(M,N,K) | TH=1 | TH=20 |
|---|---|---|
| 128 * 768 * 3072 | 4.02 ms | 578 us |
| 128 * 800 * 3100 | 3.01 ms | 277 us |
- Ernie
| GEMM Size(M,N,K) | TH=1 | TH=20 | comment |
|---|---|---|---|
| 128 * 768 * 3072 | 3.45 ms | 969 us | * |
| 128 * 768 * 3072 | 4.00 ms | 530 us | 多次调用 gemm |
- TF
| GEMM Size(M,N,K) | TH=1 | TH=20 |
|---|---|---|
| 128 * 768 * 3072 | 4.84 ms | 343 us |
padding 128 * 768 * 3072 > 128 * 772 * 3076 128 * 3072 * 768 > 128 * 3076 * 772 128 * 768 * 768 > 128 * 772 * 772 128 * 2304 * 768 > 128 * 2308 * 772
-
结论 相同的
Gemm操作，在单线程下Ernie和 UT相近，但在多线程下Ernie和 UT 差了一倍。 在Ernie计算Gemm 耗时占比约为35ms/50ms=70%,还算正常. -
Next
~1. 用 VTune 细看多线程区别~
~2. 关掉HT重新测试一遍(done)~
3. 用Vtune 查看 UT 和 Ernie 在 memory 上的区别
- Ernie 20 TH 是瓶颈在 L3 和 DDR,并且 和 TF走了不同的MKL path
~4. 在Ernie 中 同时跑多次Gemm, 观察Gemm 耗时~
https://github.com/PaddlePaddle/benchmark/issues/180#issuecomment-527756936 就是这里的，应该一样。
结论
Ernie 20 线程瓶颈在 L3 和 DDR， 原因是Ernie MKL 在20 线程的时候走了不同于TF 的MKL path。由于TF 中添加了Padding，如下
128 * 768 * 3072 > 128 * 772 * 3076 128 * 3072 * 768 > 128 * 3076 * 772 128 * 768 * 768 > 128 * 772 * 772 128 * 2304 * 768 > 128 * 2308 * 772
但Ernie 中没有，这个Padding 会导致Ernie 走到了另一个比较差的MKL path，解决方法是在Paddle中添加padding，我会先尝试一下。
感谢，但是关于padding我们已经尝试过了，多线程问题不在于padding与否，数据我们之前应该都已经同步过了。 首先，跟tf无关的事情是在你的单侧里面mkl的20线程现场正常，这个跟是否padding无关。
其次，根据你现在的结论，那依然解释不了为什么在paddle中多线程没问题，然而在erinie时有问题。
上面的结论只适用于 AVX512， AVX2 是正常的
这个现场很重要啊，我们可以也试下。
Ernie 20 线程瓶颈在 L3 和 DDR
这里可以麻烦再介绍详细点吗
-
1 thread: UT TF Ernie 最终走的MKL path 都一样
-
20 threads: UT 和 Ernie 走的一样, TF 走的是另一个(和1 thread 的一样)
-
20 threads: 加了padding之后 UT 可以和TF 持平. 并且走的 MKL path 和TF一致。
-
Next 我用VTune 查看 Paddle 加了padding 的branch，看是否和 UT, TF 走到了同一个path
Vtune 显示 Gaowei8 padding branch 走的和 TF 一致
耗时对比MKL VERBOSE
| size | Ernie | TF |
|---|---|---|
| 3072,128,768 | 352 us | 340 us |
| 768,128,3072 | 340 us | 338 us |
| 768,128,768 | 122 us | 141 us |
基本可以认定, 加了padding 之后 Ernie 和 TF 在 gemm 操作上耗时一致.
model 整体耗时
| Ernie | TF |
|---|---|
| 33 ms | 20ms |
@GaoWei8 麻烦更新下 在docker 中的 benchmark 数据
上图包含FC weight的申请内存, 数据拷贝 以及 释放内存和FC的输入，输出的数据拷贝。
FC，输入输出和weight的申请内存, 数据拷贝 以及 释放内存的全部时间。
可以根据下图计算得到，3.77ms，占总时间37.3145的10.1%
2019-10-14 Ernie 多线程排查报告总结
-
Ernie 多线程排查报告总结
- 背景
- 实验和结论
- 下一步
背景
简要说明:
- Ernie[1]单线程优化已经足够
- Ernie 多线程[2]性能和竞品 TF[3] 相距甚远(需要排查的问题))
同时 Intel 也在 6248上(和6148配置相似)复现了和百度相似的多线程 benchmark[4], 如下：
| OMP_NUM_THREADS | Ernie | TF |
|---|---|---|
| 20 | 55 ms | 20ms |
Ernie 的多线程性能落后竞品 TF ~3x, 复现 Intel benchmark 参考 Github readme[5]
详细信息参考 issue 180, Luotao 的总结, 以及 Bingyang 的 ERNIE BERT 分析.
实验和结论
comment 做了 M=128, K=768, N=3072 的三组实验
- 写 UT(每次计算不刷新数据) 调 Intel-MKLML 小包计算并打开 MKL_VERBOSE 收集对应维度的 Gemm 耗时
- Erine 打开 MKL_VERBOSE 收集对应维度 Gemm 耗时
- TF 打开 MKL_VERBOSE 收集对应维度 Gemm 耗时
- 以及在 UT 中对数据维度做 padding(改变数据维度)
得到的结论是:
-
UT 和 Ernie 多次调用 Gemm 的 Gemm 耗时相同, 此时 UT 比 Erine 性能好是因为 UT 所需数据都是 L3, 不需要访问 DDR.
-
UT 比 TF 慢大约1 倍
-
padding 后的 UT 性能和 TF 打平
-
Vtune 查看后发现:
- 单线程下 UT, Ernie 和 TF 在 MKL 中走同一个 path [MKL BLAS]@avx512_sgemm_kernel_nocopy_NN_b0*
- 20 线程下 UT, Ernie 走的是 [MKL BLAS]@avx512_sgemm_scopy_down48_ea, 而 TF 走的是 [MKL BLAS]@avx512_sgemm_kernel_nocopy_NN_b0.
推测原因是 因为不同的 MKL path 导致的, 而 padding 是导致了相同 Gemm 操作走不同 MKL path.
comment 测试了 添加 和 TF 相同的 padding 之后 MKL_VERBOSE 打出的 Gemm 耗时, 如下
| gemm size | Ernie | TF |
|---|---|---|
| 3072,128,768 | 352 us | 340 us |
| 768,128,3072 | 340 us | 338 us |
| 768,128,768 | 122 us | 141 us |
可以看出加上 padding 之后 MKL_VERBOSE 打出的 Gemm 耗时如下, 基本可以认为 对 Gemm 做 padding 之后, Ernie 的 Gemm 性能可以和竞品 TF 持平, 并且走的相同的 MKL path [MKL BLAS]@avx512_sgemm_kernel_nocopy_NN_b0.
同时文章[6]建议当用 MKL 做 Gemm 操作的时候, 如果 size 是 128 的整数倍, MKL_sgemm 的多线程性能急剧下降, 此时需要对Gemm 的维度做 padding 以方式性能损失.
但是加 padding 之后 model 性能测试如下, 发现 Ernie 仍旧慢于竞品 37%
| - | Ernie | TF | Performance gap |
|---|---|---|---|
| 非 docker | 33 ms | 20 ms | 65% |
| docker | 36 ms | 26 ms | 37% |
comment 测试了为了 padding 额外做的内存申请，数据拷贝以及内存释放的时间,如下
| - | Ernie | TF | Performance gap |
|---|---|---|---|
| docker | 32 ms | 26 ms | 23% |
可以看出这部分时间占总时间的 10% 也就是 3.77s, 去除掉这部分耗时可以得出 Ernie 和 TF 还有 23% 的性能差距.
下一步
- 查找在20多线程下 Ernie 比 TF 慢23% 的原因(BUG)
[1] 用 Paddle 预测库跑 BERT Inference.
[2] 多线程 benchmark 都是基于 20 线程, 并且做了 numactl 绑定 cpu 核心.
[3] 在 tensorflow 将 BERT 所有 OP 都 fused 起来并且对 Gemm 做 padding.
[4] Intel 所有 benchmark 都是在 Intel(R) Xeon Gold 6248 2.50GHz, 关掉超线程和 Turbo Boost, 以及非 Docker 环境下测得.
[5] ERNIE benchmark, TF benchmark
[6] Tips to Measure the Performance of Matrix Multiplication Using Intel® MKL
