Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Optimize inference performance of ERNIE on CPU #180

Open
3 tasks
tensor-tang opened this issue Aug 20, 2019 · 51 comments
Open
3 tasks

Optimize inference performance of ERNIE on CPU #180

tensor-tang opened this issue Aug 20, 2019 · 51 comments
Assignees

Comments

@tensor-tang
Copy link
Contributor

tensor-tang commented Aug 20, 2019

负责人

@tensor-tang @GaoWei8

机器型号

6148

commit号

based on #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

@tensor-tang tensor-tang self-assigned this Aug 20, 2019
@tensor-tang
Copy link
Contributor Author

tensor-tang commented Aug 21, 2019

全量结果

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

@tensor-tang
Copy link
Contributor Author

单线程 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

@tensor-tang
Copy link
Contributor Author

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

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Aug 28, 2019

baseline:
在CPU上面的测试数据
@徐屹, 82.7681ms
intel BRTR 28.62ms

优化迭代:
82.7681 ms → 60.3766 ms (提升27%)
屏幕快照 2019-08-28 14 59 11

@luotao1
Copy link
Collaborator

luotao1 commented Sep 2, 2019

ERNIE BERT 分析.pdf
@bingyanghuang 的分析。

主要结论为

  • 单线程
    • 状态
      • 去掉冗余的reshape和transpose后,在BERT base上的性能提升1.8%
      • 在FP32 上进一步优化空间较少
    • 下一步
      • 使用INT8量化进一步优化FC,以达到较为大幅度的性能提升
  • 多线程
    • 状态
      • 与竞品的2倍左右的差距初步怀疑是paddle多线程中存在一些bug导致CPU空转时间较长
    • 下一步
      • 查看Vtune log找到paddle 多线程缓慢的root cause
      • 总结竞品所使用的优化方式

@tensor-tang
Copy link
Contributor Author

一个需求:

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

@tensor-tang
Copy link
Contributor Author

tensor-tang commented Sep 4, 2019

Next step

1. mkl sgemm 尺寸测试

在ernie中

image

在paddle中单独测试,包含了每次iteration刷新矩阵X。

image

2. 去掉锁。PaddlePaddle/Paddle#19624

3. 更优可以layernorm的多线程问题。

@zhaify
Copy link

zhaify commented Sep 11, 2019

@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 耗时

@tensor-tang
Copy link
Contributor Author

paddle的单侧数据我们也测了,20threads 时是600us左右,貌似跟你这里的结果不一致呢。

@zhaify
Copy link

zhaify commented Sep 11, 2019

paddle的单侧数据我们也测了,20threads 时是600us左右,貌似跟你这里的结果不一致呢。

UT 单测和我们的数据一样吗?

@tensor-tang
Copy link
Contributor Author

#180 (comment) 就是这里的,应该一样。

@zhaify
Copy link

zhaify commented Sep 11, 2019

我更新了128 * 800 * 3100, 可以看出 UT 加上Padding 可以和TF 打平。

@tensor-tang
Copy link
Contributor Author

嗯赞,那先首要一起分析下多线程问题。

@zhaify
Copy link

zhaify commented Sep 25, 2019

结论

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,我会先尝试一下。

@zhaify
Copy link

zhaify commented Sep 25, 2019

上面的结论只适用于 AVX512, AVX2 是正常的

@tensor-tang
Copy link
Contributor Author

感谢,但是关于padding我们已经尝试过了,多线程问题不在于padding与否,数据我们之前应该都已经同步过了。
首先,跟tf无关的事情是在你的单侧里面mkl的20线程现场正常,这个跟是否padding无关。

其次,根据你现在的结论,那依然解释不了为什么在paddle中多线程没问题,然而在erinie时有问题。

@tensor-tang
Copy link
Contributor Author

上面的结论只适用于 AVX512, AVX2 是正常的

这个现场很重要啊,我们可以也试下。

Ernie 20 线程瓶颈在 L3 和 DDR

这里可以麻烦再介绍详细点吗

@zhaify
Copy link

zhaify commented Sep 25, 2019

  • 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

@zhaify
Copy link

zhaify commented Sep 26, 2019

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 数据

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Sep 26, 2019

使用numactl绑定cpu,最新 ernie 整体耗时36.17ms。
TF 目前docker内测量,耗时 26.39ms。

@zhaify
Copy link

zhaify commented Sep 26, 2019

目前的结论是 padding 之后

  • docker 环境下 Ernie 比TF 多37%
  • 非docker 环境下 Ernie 比TF 多 65%

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 11, 2019

目前Ernie与tf性能对比
屏幕快照 2019-10-11 10 02 05

@zhaify
Copy link

zhaify commented Oct 11, 2019

@GaoWei8 更新 padding 后单线程 数据

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 11, 2019

@zhaify
更新单线程加padding,numactl绑核后Ernie与tf性能对比。
屏幕快照 2019-10-11 13 09 23

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 11, 2019

@zhaify
测试padding中memory分配占用时间,仅占总时间的7%。
2c2ba034dad9834118d4bbb9c

@zhaify
Copy link

zhaify commented Oct 12, 2019

padding memory time cost 包含 申请内存, 数据拷贝 以及 释放内存 是吗?

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 12, 2019

上图包含FC weight的申请内存, 数据拷贝 以及 释放内存和FC的输入,输出的数据拷贝。
FC,输入输出和weight的申请内存, 数据拷贝 以及 释放内存的全部时间。
可以根据下图计算得到,3.77ms,占总时间37.3145的10.1%
725edc72e82956dd536011e30

@zhaify
Copy link

zhaify commented Oct 14, 2019

2019-10-14 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(改变数据维度)

得到的结论是:

  1. UT 和 Ernie 多次调用 Gemm 的 Gemm 耗时相同, 此时 UT 比 Erine 性能好是因为 UT 所需数据都是 L3, 不需要访问 DDR.

  2. UT 比 TF 慢大约1 倍

  3. padding 后的 UT 性能和 TF 打平

  4. 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

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 15, 2019

20线程目前仍存在现象:每隔一定周期,都会出现一个较大周期值48ms。
8607b8d8c0b013bf63787449f

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 17, 2019

@zhaify
Ernie各个线程时间对比。从8线程开始,Ernie的优势逆转。
image

@luotao1
Copy link
Collaborator

luotao1 commented Oct 17, 2019

padding能否 @bingyanghuang @jianhang-liu 来帮忙做一下?因为mkl内部对对线程应该能有更好的控制,可能性能会好一些。同时外面的多线程跟mkl内部的多线程策略可能不一样,会影响性能。

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 18, 2019

@zhaify
统计20线程中占用总时间超过1%的OP在多个线程的加速比。
5个OP(layernorm, Elementwise_add, matmul, softmax, scale) 20线程占比加起来有24.4%
有三个关键点:

  1. Layernorm是单线程OP,随着线程数的增加,性能反而更差,最终在20线程占据了10%的时间。(红色框标记)
  2. 所有OP随着线程数的增加,加速比的增长趋势放缓。
  3. Elementwise_add, matmul, softmax, scale四个OP随着线程数的增加,性能反而会下降。
    例如Elementwise_add, matmul从16线程开始,性能下降(蓝色框标记)。
    softmax, scale从20线程开始,性能下降(蓝色框标记)。
    f525b960f7a06de66a3adf34f

5个OP在单线程中时间占比
Layernorm 1.39%
Elementwise_add 3.83%
matmul 2.63%
softmax 0.998%
scale 0.003%

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 18, 2019

单线程mkldnn

关掉MKLDNN对单线程的影响

  1. transpose2, softmax增长倍数较大,这两个OP使用了mkldnn库
  2. Elementwise_add, Layernorm 消耗时间减少

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 18, 2019

@zhaify
20线程mkldnn
关掉MKLDNN对20线程的影响

  1. transpose2, softmax增长倍数较大,这两个OP使用了mkldnn库。
  2. Layernorm 消耗时间减少。原因?
  3. FC,Elementwise_add, matmul和scale三个OP时间消耗都有较大比例的增加。

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 21, 2019

@luotao1 @bingyanghuang
下一步可以优化的方向:

  1. mkl内部加padding
  2. Layernorm多线程
  3. 使用Vtune查看多线程中OP实际时间占比。

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 22, 2019

@bingyanghuang
测试仅部分OP打开MKLDNN开关
单线程耗时降低 1.93%
20线程耗时增加 5.74%
MKLDNN部分OP

具体OP分析:(主要关注耗时增加的OP)
MKLDNN 单线程 部分OP
耗时增加超过1%的OP:elementwise_add,softmax
耗时减少OP:Layernorm

MKLDNN 20线程 部分OP
耗时增加超过1%OP:elementwise_add,softmax,transpose2
耗时减少OP:Layernorm,matmul,scale

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 23, 2019

@bingyanghuang
Vtune 最新20线程收集结果(绑核,repeat =100)
命令:numactl --cpunodebind=0 --membind=0 ./run.sh 20 1 0
结论:消耗时间较大的部分仍然在MKL BLAS avx512

vtune-summary

Vtune-max

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 23, 2019

@bingyanghuang
去掉Transpose2 OP的时间变化。PR20770
PR20770

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 23, 2019

@bingyanghuang
去掉Transpose2之后只开启softmax的MKLDNN
结论:变化趋势和保持原有OP相同,单线程时间降低,20线程时间增加。
取掉transpose,mkldnn

具体OP时间变化趋势
结论:单线程和20线程的elementwise_add 和layernorm耗时降低
但是20线程的matmul,scale,softmax增加较多。
Tranpose2_MKLDNN_1

20线程
Tranpose2_MKLDNN_20

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 24, 2019

@bingyanghuang
在padding的基础上 Fuse fc+reshape+transpose+scale PR19526
现象:

  1. 做Fuse之后,加速比较初始图增加。
  2. 8线程,12线程,16线程的加速比增加趋势变大。
    怀疑:数据转换没有多线程功能,fuse会影响数据转换和多线程对数据的分配。
    Fuse加速比的增长趋势

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 25, 2019

@bingyanghuang
10月24号最新性能情况
屏幕快照 2019-10-25 15 13 18

@bingyanghuang
Copy link
Contributor

bingyanghuang commented Oct 25, 2019

Intel ARs:

  • 单线程:
  • 多线程:
    • 调查多线程spin time的具体原因 (10月30日 周三)
    • 确定matmul多线程下是否存在问题和fuse是否会影响多线程性能 (10月30日 周三)
    • 提出layernorm多线程的PR(11月1日 周五)
  • INT8:
    • 测试第一版INT8真实跑在int8 kernel上的性能,给出baseline(10月30日 周三)

Baidu ARs

  • Ernie-base 集成进CE监控。
  • Ernie-large在mkldnn下出错,给出复现环境。

@zhaify
Copy link

zhaify commented Oct 25, 2019

我们分别对比了 Erinie 和 TF 在单线程和多线程下 GEMM 耗时,结果如下:

  1. TF 把 QKV 三个和成了一个大GEMM,因此可以看出

image

  • 单线程,三个小GEMM 性能好于一个大GEMM(4310 >1260+1250+1260),解释了单线程 Ernie 性能优于TF
  • 多线程,三个小GEMM 性能慢于一个大 GEMM(277< 140+135+130 ), 最终导致总体差距在10%

image
2. Layernorm 大约10%
3. padding memory copy 大约 7%

因此可以初步得出 Ernie 下一步优化集中三部分:

  1. QKV 合并
  2. Layernorm 并行化
  3. padding 如何添加 Add padding to FC op  Paddle#19996

@luotao1 @GaoWei8

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 28, 2019

Memory cost time:
Input memory copy 占6.63%。
X1,Y1申请内存占1.46%
92cfbe5b2b68343ee3ca716d8
15bd478d6756d5dcc7e6e5e39
ea1726a785efab96d48778bf4
结果:
634f196c8fbd51e8a70e31751

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Oct 31, 2019

测试layernorm多线程结果#20895
ERNIE基准线是padding,绑核。

线程数 Ernie Layernorm多线程 差别
单线程 251.734 ms 252.809ms 增加0.43%
20线程 36.17ms 32.9599ms 优化8.88%

layernorm OP 耗时

线程数 Layernorm耗时 Layernorm多线程耗时 差别
单线程 3510.71ms 3871.79ms 增加10.29%
20线程 3853.87ms 887.446ms 减少77.00%

profile结果:
20线程:
优化前:

------------------------->     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                 74000       25284.9     0.023084    12.4887     0.341687    0.683151
thread0::layer_norm         25000       3937.93     0.130232    8.42511     0.157517    0.106396
thread0::elementwise_add    38000       2219.51     0.034009    8.85088     0.0584082   0.0599671
thread0::transpose2         48000       1585.09     0.025542    8.20309     0.0330227   0.0428262
thread0::matmul             25000       1545.47     0.022775    2.83107     0.0618188   0.0417558
thread0::scale              14000       782.073     0.005168    0.372797    0.0558624   0.0211302
thread0::softmax            12000       566.079     0.03742     8.67642     0.0471733   0.0152944
thread0::reshape2           48000       369.02      0.005627    0.381216    0.00768792  0.00997026
thread0::load               202         309.933     0.009914    179.667     1.53432     0.00837382
thread0::lookup_table       3000        266.646     0.071252    8.68299     0.0888819   0.00720428
thread0::stack              1000        74.7134     0.061754    0.132208    0.0747134   0.00201862
thread0::tanh               1000        42.7872     0.03601     2.42003     0.0427872   0.00115603
thread0::feed               4000        11.0731     0.001147    0.011736    0.00276828  0.000299176
thread0::slice              1000        10.8126     0.00959     0.03447     0.0108126   0.000292137
thread0::fetch              1000        6.10207     0.004864    0.02858     0.00610207  0.000164867

优化后:

------------------------->     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                 74000       24431.4     0.022329    10.0122     0.330154    0.740733
thread0::elementwise_add    38000       2194.6      0.033824    2.43725     0.0577527   0.0665378
thread0::transpose2         48000       1594.83     0.025024    8.91892     0.0332257   0.0483535
thread0::matmul             25000       1510.47     0.021713    2.56695     0.0604189   0.0457958
thread0::layer_norm         25000       894.7       0.03007     2.18405     0.035788    0.0271263
thread0::scale              14000       765.974     0.004714    0.567349    0.0547124   0.0232235
thread0::softmax            12000       539.04      0.037764    0.522655    0.04492     0.0163431
thread0::reshape2           48000       357.825     0.005703    0.729517    0.00745469  0.0108489
thread0::load               202         288.746     0.010017    181.046     1.42944     0.00875446
thread0::lookup_table       3000        257.236     0.071966    7.88935     0.0857453   0.0077991
thread0::stack              1000        70.5179     0.064874    0.133298    0.0705179   0.00213802
thread0::tanh               1000        48.1622     0.033202    8.11843     0.0481622   0.00146022
thread0::slice              1000        11.4222     0.010438    0.034986    0.0114222   0.000346309
thread0::feed               4000        10.9634     0.001066    0.022654    0.00274084  0.000332397
thread0::fetch              1000        6.84497     0.005802    0.024523    0.00684497  0.000207532

@bingyanghuang
Copy link
Contributor

基于现在情况的总结和大fusion的建议。
ERNIE Analysis.pdf

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Nov 20, 2019

@bingyanghuang
现象
TF20线程(10条数据,repeat=100)的数据波动过大(26~29ms)。需要修改TF的测试环境,给出准确的竞品数据。

实验&分析
初始条件:初始TF测试环境为随机输入数据(placeholder_0),长度为(1,128)且数据量为10条。

实验:

  1. 将数据数量扩大成5010条,repeat=1,但仍为随机输入数据,耗时30.1767ms。
  2. 数据量5010条,repeat=1,将TF输入数据改成和paddle相同的输入数据, 耗时55.152522 ms。
  3. 数据量5010条,repeat=1,将paddle中的输入数据为0的数据替换成1, 耗时29.954831 ms。

疑惑:造成数据0,1耗时相差过大的原因,

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Nov 20, 2019

全数据集数据
padding weight pass优化 #20972 后,5010个数据,repeat=1(TF暂时使用随机数据):
单线程: 251.464 ms(paddle)VS 292.43 ms(TF)
20线程:29.8818(paddle)VS 30.1767 ms(TF)

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Nov 27, 2019

关于统一数据集和@bingyanghuang的讨论:
主要是两个点:

  1. 5010条数据,repeat=1,这个配置是否合理。
  2. TF输入数据使用随机数据是否合理。

初步结论:

  1. 可以多次实验,paddle和TF得到的数据都较为稳定即可。
  2. tf的模型是Intel内部的一个优化模型demo,并非实际用户使用模型。因此数据的输入在初始设计时并没有实际意义,所以0和1没有实际数据的区分度。并且tf模型是使用1条128长度的随机数据,采用一个lookup_table的近似功能模块得到一个长度为768的数据。因此按照tf环境的现有模型没办法完成数据集的对齐。

@GaoWei8
Copy link
Collaborator

GaoWei8 commented Jan 15, 2020

@bingyanghuang
更新Paddle与TF性能数据:
未绑核数据:

线程数 Ernie TF 差别
单线程 252.257 ms 290.192ms 优化13.07%
20线程 32.4479ms 32.9599ms 打平

绑核
命令:numactl --cpunodebind=0 --membind=0
数据:

线程数 Ernie TF 差别
单线程 250.615 253.618670 优化1.18%
20线程 26.7276 21.842005 差别22.37%

@luotao1
Copy link
Collaborator

luotao1 commented Jan 15, 2020

@GaoWei8 请补充下绑核的命令

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
None yet
Projects
None yet
Development

No branches or pull requests

5 participants