Skip to content
体验新版
项目
组织
正在加载...
登录
切换导航
打开侧边栏
Xiaomi
Mace
提交
3e3b214b
Mace
项目概览
Xiaomi
/
Mace
通知
106
Star
40
Fork
27
代码
文件
提交
分支
Tags
贡献者
分支图
Diff
Issue
0
列表
看板
标记
里程碑
合并请求
0
DevOps
流水线
流水线任务
计划
Wiki
0
Wiki
分析
仓库
DevOps
项目成员
Pages
Mace
项目概览
项目概览
详情
发布
仓库
仓库
文件
提交
分支
标签
贡献者
分支图
比较
Issue
0
Issue
0
列表
看板
标记
里程碑
合并请求
0
合并请求
0
Pages
DevOps
DevOps
流水线
流水线任务
计划
分析
分析
仓库分析
DevOps
Wiki
0
Wiki
成员
成员
收起侧边栏
关闭侧边栏
动态
分支图
创建新Issue
流水线任务
提交
Issue看板
体验新版 GitCode,发现更多精彩内容 >>
提交
3e3b214b
编写于
3月 01, 2018
作者:
L
Liangliang He
浏览文件
操作
浏览文件
下载
电子邮件补丁
差异文件
Fix softmax performance issue
上级
3236ad7c
变更
3
隐藏空白更改
内联
并排
Showing
3 changed file
with
37 addition
and
38 deletion
+37
-38
mace/kernels/neon/conv_2d_neon_1x1.cc
mace/kernels/neon/conv_2d_neon_1x1.cc
+0
-2
mace/kernels/softmax.h
mace/kernels/softmax.h
+28
-28
mace/ops/softmax_benchmark.cc
mace/ops/softmax_benchmark.cc
+9
-8
未找到文件。
mace/kernels/neon/conv_2d_neon_1x1.cc
浏览文件 @
3e3b214b
...
...
@@ -264,7 +264,6 @@ void Conv2dNeonK1x1S1(const float *input, // NCHW
bias
?
bias
[
i
]
:
0
);
}
}
// benchmark omp collapsed(2)
#pragma omp parallel for collapse(2)
for
(
index_t
n
=
0
;
n
<
batch
;
++
n
)
{
for
(
index_t
c
=
0
;
c
<
round_up_channels
;
c
+=
kOutputChannelBlockSize
)
{
...
...
@@ -326,7 +325,6 @@ void Conv2dNeonPixelK1x1S1(
const
index_t
total_loops
=
total_pixels
>>
3
;
const
index_t
loop_remaining
=
total_pixels
&
7
;
// benchmark omp collapsed(2)
#pragma omp parallel for collapse(2)
for
(
index_t
n
=
0
;
n
<
batch
;
++
n
)
{
for
(
index_t
c
=
0
;
c
<
channels
;
++
c
)
{
...
...
mace/kernels/softmax.h
浏览文件 @
3e3b214b
...
...
@@ -6,55 +6,55 @@
#define MACE_KERNELS_SOFTMAX_H_
#include "mace/core/future.h"
#include "mace/core/runtime/opencl/cl2_header.h"
#include "mace/core/tensor.h"
#include "mace/public/mace.h"
#include "mace/
core/runtime/opencl/cl2_header
.h"
#include "mace/
utils/utils
.h"
namespace
mace
{
namespace
kernels
{
template
<
DeviceType
D
,
typename
T
>
struct
SoftmaxFunctor
{
void
operator
()(
const
Tensor
*
logits
,
Tensor
*
output
,
StatsFuture
*
future
)
{
void
operator
()(
const
Tensor
*
logits
,
Tensor
*
output
,
StatsFuture
*
future
)
{
Tensor
::
MappingGuard
logits_guard
(
logits
);
Tensor
::
MappingGuard
output_guard
(
output
);
const
T
*
logits_ptr
=
logits
->
data
<
T
>
();
T
*
output_ptr
=
output
->
mutable_data
<
T
>
();
auto
&
logits_shape
=
logits
->
shape
();
const
index_t
batch_size
=
std
::
accumulate
(
logits_shape
.
begin
(),
logits_shape
.
end
()
-
1
,
1
,
std
::
multiplies
<
index_t
>
());
const
index_t
batch_size
=
std
::
accumulate
(
logits_shape
.
begin
(),
logits_shape
.
end
()
-
1
,
1
,
std
::
multiplies
<
index_t
>
());
const
index_t
num_classes
=
logits_shape
.
back
();
#pragma omp parallel for
for
(
index_t
i
=
0
;
i
<
batch_size
;
++
i
)
{
const
index_t
pos
=
i
*
num_classes
;
T
max_value
=
logits_ptr
[
pos
];
for
(
index_t
c
=
1
;
c
<
num_classes
;
++
c
)
{
max_value
=
std
::
max
(
max_value
,
logits_ptr
[
pos
+
c
]);
}
// TODO: check overflow?
T
sum
=
0
;
#pragma omp parallel
{
// Allocate per thread buffer
std
::
vector
<
T
>
exp_data
(
num_classes
);
for
(
index_t
c
=
0
;
c
<
num_classes
;
++
c
)
{
exp_data
[
c
]
=
::
exp
((
logits_ptr
[
pos
+
c
]
-
max_value
));
sum
+=
exp_data
[
c
];
}
for
(
index_t
c
=
0
;
c
<
num_classes
;
++
c
)
{
output_ptr
[
pos
+
c
]
=
exp_data
[
c
]
/
sum
;
#pragma omp for
for
(
index_t
i
=
0
;
i
<
batch_size
;
++
i
)
{
const
index_t
pos
=
i
*
num_classes
;
T
max_value
=
logits_ptr
[
pos
];
for
(
index_t
c
=
1
;
c
<
num_classes
;
++
c
)
{
max_value
=
std
::
max
(
max_value
,
logits_ptr
[
pos
+
c
]);
}
// TODO: check overflow?
T
sum
=
0
;
for
(
index_t
c
=
0
;
c
<
num_classes
;
++
c
)
{
exp_data
[
c
]
=
::
exp
((
logits_ptr
[
pos
+
c
]
-
max_value
));
sum
+=
exp_data
[
c
];
}
for
(
index_t
c
=
0
;
c
<
num_classes
;
++
c
)
{
output_ptr
[
pos
+
c
]
=
exp_data
[
c
]
/
sum
;
}
}
}
}
};
template
<
typename
T
>
template
<
typename
T
>
struct
SoftmaxFunctor
<
DeviceType
::
OPENCL
,
T
>
{
void
operator
()(
const
Tensor
*
logits
,
Tensor
*
output
,
StatsFuture
*
future
);
void
operator
()(
const
Tensor
*
logits
,
Tensor
*
output
,
StatsFuture
*
future
);
cl
::
Kernel
kernel_
;
};
...
...
mace/ops/softmax_benchmark.cc
浏览文件 @
3e3b214b
...
...
@@ -46,12 +46,13 @@ static void SoftmaxBenchmark(
net
.
Sync
();
}
#define BM_SOFTMAX_MACRO(N, C, H, W, TYPE, DEVICE) \
static void BM_SOFTMAX_##N##_##C##_##H##_##W##_##TYPE##_##DEVICE(int iters) { \
#define BM_SOFTMAX_MACRO(N, C, H, W, TYPE, DEVICE) \
static void BM_SOFTMAX_##N##_##C##_##H##_##W##_##TYPE##_##DEVICE( \
int iters) { \
const int64_t tot = static_cast<int64_t>(iters) * N * C * H * W; \
mace::testing::MaccProcessed(tot); \
mace::testing::MaccProcessed(tot);
\
mace::testing::BytesProcessed(tot *(sizeof(TYPE))); \
SoftmaxBenchmark<DEVICE, TYPE>(iters, N, C, H, W);
\
SoftmaxBenchmark<DEVICE, TYPE>(iters, N, C, H, W); \
} \
BENCHMARK(BM_SOFTMAX_##N##_##C##_##H##_##W##_##TYPE##_##DEVICE)
...
...
@@ -60,9 +61,9 @@ static void SoftmaxBenchmark(
BM_SOFTMAX_MACRO(N, C, H, W, float, OPENCL); \
BM_SOFTMAX_MACRO(N, C, H, W, half, OPENCL);
BM_SOFTMAX
(
1
,
1
,
512
,
512
);
BM_SOFTMAX
(
1
,
3
,
128
,
128
);
BM_SOFTMAX
(
1
,
2
,
512
,
512
);
BM_SOFTMAX
(
1
,
3
,
512
,
512
);
BM_SOFTMAX
(
1
,
32
,
112
,
112
);
BM_SOFTMAX
(
1
,
64
,
256
,
256
);
BM_SOFTMAX
(
1
,
4
,
512
,
512
);
BM_SOFTMAX
(
1
,
10
,
256
,
256
);
BM_SOFTMAX
(
1
,
1024
,
7
,
7
);
}
// namespace mace
编辑
预览
Markdown
is supported
0%
请重试
或
添加新附件
.
添加附件
取消
You are about to add
0
people
to the discussion. Proceed with caution.
先完成此消息的编辑!
取消
想要评论请
注册
或
登录