Skip to content
GitLab
Menu
Projects
Groups
Snippets
Loading...
Help
Help
Support
Community forum
Keyboard shortcuts
?
Submit feedback
Contribute to GitLab
Sign in / Register
Toggle navigation
Menu
Open sidebar
gaoqiong
MIGraphX
Commits
4c638f13
Unverified
Commit
4c638f13
authored
Nov 02, 2018
by
Paul Fultz II
Committed by
GitHub
Nov 02, 2018
Browse files
Merge branch 'master' into globalavgpool
parents
856fc58a
0d0778b7
Changes
11
Hide whitespace changes
Inline
Side-by-side
Showing
11 changed files
with
75 additions
and
57 deletions
+75
-57
doc/src/reference/targets.rst
doc/src/reference/targets.rst
+1
-1
src/onnx/cifar10.cpp
src/onnx/cifar10.cpp
+2
-2
src/onnx/verify_onnx.cpp
src/onnx/verify_onnx.cpp
+2
-2
src/targets/cpu/CMakeLists.txt
src/targets/cpu/CMakeLists.txt
+2
-2
src/targets/cpu/include/migraph/cpu/lowering.hpp
src/targets/cpu/include/migraph/cpu/lowering.hpp
+1
-1
src/targets/cpu/include/migraph/cpu/target.hpp
src/targets/cpu/include/migraph/cpu/target.hpp
+1
-1
src/targets/cpu/lowering.cpp
src/targets/cpu/lowering.cpp
+2
-2
src/targets/cpu/target.cpp
src/targets/cpu/target.cpp
+18
-0
test/cpu_ops_test.cpp
test/cpu_ops_test.cpp
+41
-41
test/fwd_conv_batchnorm_rewrite_test.cpp
test/fwd_conv_batchnorm_rewrite_test.cpp
+3
-3
test/gpu/miopen.cpp
test/gpu/miopen.cpp
+2
-2
No files found.
doc/src/reference/targets.rst
View file @
4c638f13
...
...
@@ -14,5 +14,5 @@ gpu::target
cpu::target
-----------
.. doxygenstruct:: migraph::cpu::
cpu_
target
.. doxygenstruct:: migraph::cpu::target
src/onnx/cifar10.cpp
View file @
4c638f13
...
...
@@ -6,7 +6,7 @@
#include <migraph/onnx.hpp>
#include <migraph/cpu/
cpu_
target.hpp>
#include <migraph/cpu/target.hpp>
#include <migraph/gpu/target.hpp>
#include <migraph/gpu/hip.hpp>
#include <migraph/generate.hpp>
...
...
@@ -86,7 +86,7 @@ int main(int argc, char const* argv[])
else
{
// CPU target
prog
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
prog
.
compile
(
migraph
::
cpu
::
target
{});
auto
s
=
migraph
::
shape
{
migraph
::
shape
::
float_type
,
{
1
,
3
,
32
,
32
}};
auto
labels
=
imageset
.
first
;
auto
input
=
imageset
.
second
;
...
...
src/onnx/verify_onnx.cpp
View file @
4c638f13
#include <migraph/onnx.hpp>
#include <migraph/cpu/
cpu_
target.hpp>
#include <migraph/cpu/target.hpp>
#include <migraph/gpu/target.hpp>
#include <migraph/gpu/hip.hpp>
#include <migraph/generate.hpp>
...
...
@@ -18,7 +18,7 @@ template <class F>
migraph
::
argument
run_cpu
(
F
f
)
{
auto
p
=
f
();
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
migraph
::
program
::
parameter_map
m
;
for
(
auto
&&
x
:
p
.
get_parameter_shapes
())
{
...
...
src/targets/cpu/CMakeLists.txt
View file @
4c638f13
add_library
(
migraph_cpu
cpu_
target.cpp
cpu_
lowering.cpp
target.cpp
lowering.cpp
gemm.cpp
)
...
...
src/targets/cpu/include/migraph/cpu/
cpu_
lowering.hpp
→
src/targets/cpu/include/migraph/cpu/lowering.hpp
View file @
4c638f13
...
...
@@ -6,7 +6,7 @@
namespace
migraph
{
namespace
cpu
{
struct
cpu_
lowering
struct
lowering
{
std
::
string
name
()
const
{
return
"cpu::lowering"
;
}
void
apply
(
program
&
p
)
const
;
...
...
src/targets/cpu/include/migraph/cpu/
cpu_
target.hpp
→
src/targets/cpu/include/migraph/cpu/target.hpp
View file @
4c638f13
...
...
@@ -7,7 +7,7 @@
namespace
migraph
{
namespace
cpu
{
struct
cpu_
target
struct
target
{
std
::
string
name
()
const
;
std
::
vector
<
pass
>
get_passes
(
migraph
::
context
&
ctx
)
const
;
...
...
src/targets/cpu/
cpu_
lowering.cpp
→
src/targets/cpu/lowering.cpp
View file @
4c638f13
#include <migraph/cpu/
cpu_
lowering.hpp>
#include <migraph/cpu/lowering.hpp>
#include <migraph/instruction.hpp>
#include <migraph/dfor.hpp>
#include <migraph/operators.hpp>
...
...
@@ -664,7 +664,7 @@ struct cpu_apply
}
};
void
cpu_
lowering
::
apply
(
program
&
p
)
const
{
cpu_apply
{
&
p
}.
apply
();
}
void
lowering
::
apply
(
program
&
p
)
const
{
cpu_apply
{
&
p
}.
apply
();
}
}
// namespace cpu
...
...
src/targets/cpu/
cpu_
target.cpp
→
src/targets/cpu/target.cpp
View file @
4c638f13
#include <migraph/cpu/
cpu_
target.hpp>
#include <migraph/cpu/
cpu_
lowering.hpp>
#include <migraph/cpu/target.hpp>
#include <migraph/cpu/lowering.hpp>
#include <migraph/auto_contiguous.hpp>
namespace
migraph
{
namespace
cpu
{
std
::
string
cpu_
target
::
name
()
const
{
return
"cpu"
;
}
std
::
string
target
::
name
()
const
{
return
"cpu"
;
}
std
::
vector
<
pass
>
cpu_
target
::
get_passes
(
migraph
::
context
&
)
const
std
::
vector
<
pass
>
target
::
get_passes
(
migraph
::
context
&
)
const
{
return
{
auto_contiguous
{},
cpu_
lowering
{}};
return
{
auto_contiguous
{},
lowering
{}};
}
}
// namespace cpu
...
...
test/cpu_ops_test.cpp
View file @
4c638f13
...
...
@@ -3,7 +3,7 @@
#include <migraph/literal.hpp>
#include <migraph/operators.hpp>
#include <migraph/instruction.hpp>
#include <migraph/cpu/
cpu_
target.hpp>
#include <migraph/cpu/target.hpp>
#include <migraph/verify.hpp>
#include "test.hpp"
...
...
@@ -18,7 +18,7 @@ void slice_test()
p
.
add_instruction
(
migraph
::
op
::
slice
{{
2
},
{
1
},
{
3
}},
l0
);
migraph
::
shape
s2
{
migraph
::
shape
::
int32_type
,
{
2
,
2
,
2
},
{
6
,
3
,
1
}};
EXPECT
(
p
.
get_shape
()
==
s2
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
migraph
::
shape
sresult
{
migraph
::
shape
::
int32_type
,
{
2
,
2
,
2
},
{
4
,
2
,
1
}};
auto
result
=
p
.
eval
({});
std
::
vector
<
int
>
gold
=
{
1
,
2
,
4
,
5
,
7
,
8
,
10
,
11
};
...
...
@@ -36,7 +36,7 @@ void slice_test()
p
.
add_instruction
(
migraph
::
op
::
slice
{{
0
,
1
,
2
},
{
0
,
0
,
0
},
{
2
,
2
,
2
}},
l0
);
migraph
::
shape
s2
{
migraph
::
shape
::
int32_type
,
{
2
,
2
,
2
},
{
6
,
3
,
1
}};
EXPECT
(
p
.
get_shape
()
==
s2
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
migraph
::
shape
sresult
{
migraph
::
shape
::
int32_type
,
{
2
,
2
,
2
},
{
4
,
2
,
1
}};
auto
result
=
p
.
eval
({});
std
::
vector
<
int
>
gold
=
{
0
,
1
,
3
,
4
,
6
,
7
,
9
,
10
};
...
...
@@ -62,7 +62,7 @@ void concat_test()
auto
l1
=
p
.
add_literal
(
migraph
::
literal
{
s1
,
data1
});
auto
l2
=
p
.
add_literal
(
migraph
::
literal
{
s2
,
data2
});
p
.
add_instruction
(
migraph
::
op
::
concat
{
axis
},
l0
,
l1
,
l2
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
int
>
gold
=
{
0
,
1
,
2
,
3
,
4
,
10
,
5
,
6
,
7
,
8
,
9
,
20
};
std
::
vector
<
int
>
results_vector
(
2
*
6
);
...
...
@@ -85,7 +85,7 @@ void concat_test()
auto
l1
=
p
.
add_literal
(
migraph
::
literal
{
s1
,
data1
});
auto
l2
=
p
.
add_literal
(
migraph
::
literal
{
s2
,
data2
});
p
.
add_instruction
(
migraph
::
op
::
concat
{
axis
},
l0
,
l1
,
l2
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
int
>
gold
=
{
0
,
1
,
2
,
3
,
4
,
5
,
6
,
7
,
8
,
9
,
10
,
11
};
std
::
vector
<
int
>
results_vector
(
6
*
2
);
...
...
@@ -106,7 +106,7 @@ void squeeze_test()
migraph
::
shape
s2
{
migraph
::
shape
::
float_type
,
{
4
,
3
,
1
,
3
}};
auto
l0
=
p
.
add_literal
(
migraph
::
literal
{
s1
,
data
});
p
.
add_instruction
(
migraph
::
op
::
squeeze
{{
1
}},
l0
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
EXPECT
(
result
.
get_shape
()
==
s2
);
}
...
...
@@ -117,7 +117,7 @@ void squeeze_test()
migraph
::
shape
s2
{
migraph
::
shape
::
float_type
,
{
4
,
1
,
3
,
3
}};
auto
l0
=
p
.
add_literal
(
migraph
::
literal
{
s1
,
data
});
p
.
add_instruction
(
migraph
::
op
::
squeeze
{{
3
}},
l0
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
EXPECT
(
result
.
get_shape
()
==
s2
);
}
...
...
@@ -128,7 +128,7 @@ void squeeze_test()
migraph
::
shape
s2
{
migraph
::
shape
::
float_type
,
{
4
,
3
,
3
}};
auto
l0
=
p
.
add_literal
(
migraph
::
literal
{
s1
,
data
});
p
.
add_instruction
(
migraph
::
op
::
squeeze
{},
l0
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
EXPECT
(
result
.
get_shape
()
==
s2
);
}
...
...
@@ -143,7 +143,7 @@ void unsqueeze_test()
migraph
::
shape
s2
{
migraph
::
shape
::
float_type
,
{
4
,
1
,
3
,
3
}};
auto
l0
=
p
.
add_literal
(
migraph
::
literal
{
s1
,
data
});
p
.
add_instruction
(
migraph
::
op
::
unsqueeze
{{
1
}},
l0
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
EXPECT
(
result
.
get_shape
()
==
s2
);
}
...
...
@@ -154,7 +154,7 @@ void unsqueeze_test()
migraph
::
shape
s2
{
migraph
::
shape
::
float_type
,
{
4
,
3
,
1
,
3
}};
auto
l0
=
p
.
add_literal
(
migraph
::
literal
{
s1
,
data
});
p
.
add_instruction
(
migraph
::
op
::
unsqueeze
{{
2
}},
l0
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
EXPECT
(
result
.
get_shape
()
==
s2
);
}
...
...
@@ -219,7 +219,7 @@ void im2col_3x3_no_pad_identity_test()
auto
l_image
=
p
.
add_literal
(
migraph
::
literal
{
s_image
,
input
});
auto
l_weights
=
p
.
add_literal
(
migraph
::
literal
{
s_weights
,
weights
});
p
.
add_instruction
(
migraph
::
op
::
im2col
{
padding
,
stride
,
dilation
},
l_image
,
l_weights
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
size_t
col_height
=
(
size
[
0
]
-
f
[
0
]
+
2
*
padding
[
0
])
/
stride
[
0
]
+
1
;
...
...
@@ -248,7 +248,7 @@ void im2col_3x3_no_pad_test()
auto
l_image
=
p
.
add_literal
(
migraph
::
literal
{
s_image
,
input
});
auto
l_weights
=
p
.
add_literal
(
migraph
::
literal
{
s_weights
,
weights
});
p
.
add_instruction
(
migraph
::
op
::
im2col
{
padding
,
stride
,
dilation
},
l_image
,
l_weights
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
int
>
correct
=
{
0
,
1
,
2
,
4
,
5
,
6
,
8
,
9
,
10
,
1
,
2
,
3
,
5
,
6
,
7
,
9
,
10
,
11
,
...
...
@@ -280,7 +280,7 @@ void im2col_3x3_stride_2_no_pad_test()
auto
l_image
=
p
.
add_literal
(
migraph
::
literal
{
s_image
,
input
});
auto
l_weights
=
p
.
add_literal
(
migraph
::
literal
{
s_weights
,
weights
});
p
.
add_instruction
(
migraph
::
op
::
im2col
{
padding
,
stride
,
dilation
},
l_image
,
l_weights
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
int
>
correct
=
{
0
,
1
,
2
,
6
,
7
,
8
,
12
,
13
,
14
,
2
,
3
,
4
,
...
...
@@ -313,7 +313,7 @@ void im2col_3x3_with_padding_test()
auto
l_image
=
p
.
add_literal
(
migraph
::
literal
{
s_image
,
input
});
auto
l_weights
=
p
.
add_literal
(
migraph
::
literal
{
s_weights
,
weights
});
p
.
add_instruction
(
migraph
::
op
::
im2col
{
padding
,
stride
,
dilation
},
l_image
,
l_weights
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
int
>
correct
=
{
0
,
0
,
0
,
0
,
0
,
1
,
0
,
2
,
3
,
0
,
0
,
0
,
0
,
1
,
0
,
2
,
3
,
0
,
...
...
@@ -355,7 +355,7 @@ void batch_norm_inference_test()
auto
variance
=
p
.
add_literal
(
migraph
::
literal
{
vars
,
variance_data
});
p
.
add_instruction
(
migraph
::
op
::
batch_norm_inference
{},
x
,
scale
,
bias
,
mean
,
variance
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
result_vector
(
width
*
height
*
channels
*
batches
);
...
...
@@ -385,7 +385,7 @@ void im2col_3x3_with_channels_identity_test()
auto
l_image
=
p
.
add_literal
(
migraph
::
literal
{
s_image
,
input
});
auto
l_weights
=
p
.
add_literal
(
migraph
::
literal
{
s_weights
,
weights
});
p
.
add_instruction
(
migraph
::
op
::
im2col
{
padding
,
stride
,
dilation
},
l_image
,
l_weights
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
size_t
col_height
=
(
size
[
0
]
-
f
[
0
]
+
2
*
padding
[
0
])
/
stride
[
0
]
+
1
;
...
...
@@ -401,7 +401,7 @@ void exp_test()
migraph
::
shape
s
{
migraph
::
shape
::
float_type
,
{
3
}};
auto
l
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
-
1
,
0
,
1
}});
p
.
add_instruction
(
migraph
::
op
::
exp
{},
l
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -415,7 +415,7 @@ void sin_test()
migraph
::
shape
s
{
migraph
::
shape
::
float_type
,
{
3
}};
auto
l
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
-
1
,
0
,
1
}});
p
.
add_instruction
(
migraph
::
op
::
sin
{},
l
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -429,7 +429,7 @@ void cos_test()
migraph
::
shape
s
{
migraph
::
shape
::
float_type
,
{
3
}};
auto
l
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
-
1
,
0
,
1
}});
p
.
add_instruction
(
migraph
::
op
::
cos
{},
l
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -443,7 +443,7 @@ void tan_test()
migraph
::
shape
s
{
migraph
::
shape
::
float_type
,
{
3
}};
auto
l
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
-
1
,
0
,
1
}});
p
.
add_instruction
(
migraph
::
op
::
tan
{},
l
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -458,7 +458,7 @@ void add_test()
auto
l1
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
-
1
,
0
,
1
}});
auto
l2
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
1
,
2
,
3
}});
p
.
add_instruction
(
migraph
::
op
::
add
{},
l1
,
l2
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -477,7 +477,7 @@ void broadcast_test()
auto
l1
=
p
.
add_literal
(
migraph
::
literal
{
a_shape
,
a_data
});
auto
l2
=
p
.
add_literal
(
migraph
::
literal
{
b_shape
,
b_data
});
p
.
add_instruction
(
migraph
::
op
::
broadcast
{
axis
,
l1
->
get_shape
()},
l2
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
auto
output
=
result
.
get
<
int32_t
>
();
EXPECT
(
output
(
0
,
0
)
==
-
2
);
...
...
@@ -497,7 +497,7 @@ void add_broadcast_test()
auto
l2
=
p
.
add_literal
(
migraph
::
literal
{
b_shape
,
b_data
});
auto
l3
=
p
.
add_instruction
(
migraph
::
op
::
broadcast
{
axis
,
l1
->
get_shape
()},
l2
);
p
.
add_instruction
(
migraph
::
op
::
add
{},
l1
,
l3
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
EXPECT
(
result
.
get_shape
().
packed
());
std
::
vector
<
float
>
results_vector
(
12
);
...
...
@@ -513,7 +513,7 @@ void sub_test()
auto
l1
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
-
1
,
0
,
1
}});
auto
l2
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
1
,
2
,
3
}});
p
.
add_instruction
(
migraph
::
op
::
sub
{},
l1
,
l2
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -528,7 +528,7 @@ void mul_test()
auto
l1
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
-
1
,
0
,
1
}});
auto
l2
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
1
,
2
,
3
}});
p
.
add_instruction
(
migraph
::
op
::
mul
{},
l1
,
l2
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -543,7 +543,7 @@ void div_test()
auto
l1
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
-
1.0
f
,
0.5
f
,
1.0
f
}});
auto
l2
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
1.0
f
,
2.0
f
,
4.0
f
}});
p
.
add_instruction
(
migraph
::
op
::
div
{},
l1
,
l2
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -557,7 +557,7 @@ void relu_test()
migraph
::
shape
s
{
migraph
::
shape
::
float_type
,
{
3
}};
auto
l
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
-
1.
f
,
0.
f
,
1.
f
}});
p
.
add_instruction
(
migraph
::
op
::
activation
{
"relu"
},
l
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -571,7 +571,7 @@ void leaky_relu_test()
migraph
::
shape
s
{
migraph
::
shape
::
float_type
,
{
3
}};
auto
l
=
p
.
add_literal
(
migraph
::
literal
{
s
,
{
-
1.
f
,
0.
f
,
1.
f
}});
p
.
add_instruction
(
migraph
::
op
::
leaky_relu
{
0.01
},
l
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -605,7 +605,7 @@ void imagescaler_test()
migraph
::
literal
{
migraph
::
shape
{
migraph
::
shape
::
float_type
,
{
3
}},
{
0.01
,
0.02
,
0.03
}});
auto
bias_bcast
=
p
.
add_instruction
(
migraph
::
op
::
broadcast
{
1
,
s
},
bias_vals
);
p
.
add_instruction
(
migraph
::
op
::
add
{},
img_scaled
,
bias_bcast
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
12
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -636,7 +636,7 @@ void reshape_test()
auto
l
=
p
.
add_literal
(
migraph
::
literal
{
a_shape
,
data
});
std
::
vector
<
int64_t
>
new_shape
=
{
8
,
3
,
1
,
1
};
p
.
add_instruction
(
migraph
::
op
::
reshape
{
new_shape
},
l
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -647,7 +647,7 @@ void reshape_test()
auto
l
=
p
.
add_literal
(
migraph
::
literal
{
a_shape
,
data
});
std
::
vector
<
int64_t
>
new_shape
=
{
1
,
3
,
4
,
2
};
p
.
add_instruction
(
migraph
::
op
::
reshape
{
new_shape
},
l
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -658,7 +658,7 @@ void reshape_test()
auto
l
=
p
.
add_literal
(
migraph
::
literal
{
a_shape
,
data
});
std
::
vector
<
int64_t
>
new_shape
=
{
1
,
3
,
4
,
2
};
p
.
add_instruction
(
migraph
::
op
::
reshape
{
new_shape
},
l
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
3
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -706,7 +706,7 @@ void gemm_test()
migraph
::
shape
b_shape
{
migraph
::
shape
::
get_type
<
T
>
{},
{
5
,
3
}};
auto
bl
=
p
.
add_literal
(
migraph
::
literal
{
b_shape
,
b
});
p
.
add_instruction
(
migraph
::
op
::
dot
{},
al
,
bl
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
T
>
results_vector
(
12
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -761,7 +761,7 @@ void maxpool_test()
migraph
::
shape
a_shape
{
migraph
::
shape
::
float_type
,
{
2
,
3
,
6
,
6
}};
auto
al
=
p
.
add_literal
(
migraph
::
literal
{
a_shape
,
a
});
p
.
add_instruction
(
migraph
::
op
::
pooling
{
"max"
,
{{
0
,
0
}},
{{
2
,
2
}},
{{
3
,
2
}}},
al
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
cout
<<
result
.
get_shape
()
<<
std
::
endl
;
std
::
vector
<
float
>
results_vector
(
36
);
...
...
@@ -826,7 +826,7 @@ void softmax_test()
migraph
::
shape
a_shape
{
migraph
::
shape
::
float_type
,
{
5
,
3
,
4
,
2
}};
auto
al
=
p
.
add_literal
(
migraph
::
literal
{
a_shape
,
a
});
p
.
add_instruction
(
migraph
::
op
::
softmax
{},
al
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
120
);
result
.
visit
([
&
](
auto
output
)
{
results_vector
.
assign
(
output
.
begin
(),
output
.
end
());
});
...
...
@@ -888,7 +888,7 @@ void conv2d_test()
auto
cl
=
p
.
add_literal
(
migraph
::
literal
{
c_shape
,
c
});
p
.
add_instruction
(
migraph
::
op
::
convolution
{},
al
,
cl
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
16
);
...
...
@@ -944,7 +944,7 @@ void conv2d_padding_test()
auto
cl
=
p
.
add_literal
(
migraph
::
literal
{
c_shape
,
c
});
p
.
add_instruction
(
migraph
::
op
::
convolution
{{{
1
,
1
}},
{{
1
,
1
}}},
al
,
cl
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
64
);
...
...
@@ -1005,7 +1005,7 @@ void conv2d_padding_stride_test()
auto
cl
=
p
.
add_literal
(
migraph
::
literal
{
c_shape
,
c
});
p
.
add_instruction
(
migraph
::
op
::
convolution
{{{
1
,
1
}},
{{
2
,
2
}}},
al
,
cl
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
16
);
...
...
@@ -1024,7 +1024,7 @@ void transpose_test()
auto
l
=
p
.
add_literal
(
migraph
::
literal
{
a_shape
,
data
});
std
::
vector
<
int64_t
>
perm
=
{
0
,
3
,
1
,
2
};
p
.
add_instruction
(
migraph
::
op
::
transpose
{
perm
},
l
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
result
.
visit
([
&
](
auto
output
)
{
...
...
@@ -1038,7 +1038,7 @@ void transpose_test()
std
::
vector
<
int64_t
>
perm
=
{
0
,
3
,
1
,
2
};
auto
result
=
p
.
add_instruction
(
migraph
::
op
::
transpose
{
perm
},
l
);
p
.
add_instruction
(
migraph
::
op
::
contiguous
{},
result
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result2
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
12
);
...
...
@@ -1057,7 +1057,7 @@ void contiguous_test()
migraph
::
program
p
;
auto
l
=
p
.
add_literal
(
migraph
::
literal
{
a_shape
,
data
});
p
.
add_instruction
(
migraph
::
op
::
contiguous
{},
l
);
p
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p
.
compile
(
migraph
::
cpu
::
target
{});
auto
result
=
p
.
eval
({});
std
::
vector
<
float
>
results_vector
(
12
);
...
...
test/fwd_conv_batchnorm_rewrite_test.cpp
View file @
4c638f13
#include <migraph/fwd_conv_batchnorm_rewrite.hpp>
#include <migraph/program.hpp>
#include <migraph/cpu/
cpu_
target.hpp>
#include <migraph/cpu/target.hpp>
#include <migraph/operators.hpp>
#include <migraph/instruction.hpp>
#include <test.hpp>
...
...
@@ -51,8 +51,8 @@ void fwd_conv_batchnorm_rewrite_test()
migraph
::
program
p2
=
create_program
();
migraph
::
fwd_conv_batchnorm_rewrite
opt
;
opt
.
apply
(
p2
);
p1
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p2
.
compile
(
migraph
::
cpu
::
cpu_
target
{});
p1
.
compile
(
migraph
::
cpu
::
target
{});
p2
.
compile
(
migraph
::
cpu
::
target
{});
auto
result1
=
p1
.
eval
({});
auto
result2
=
p2
.
eval
({});
...
...
test/gpu/miopen.cpp
View file @
4c638f13
...
...
@@ -2,7 +2,7 @@
#include <migraph/program.hpp>
#include <migraph/operators.hpp>
#include <migraph/generate.hpp>
#include <migraph/cpu/
cpu_
target.hpp>
#include <migraph/cpu/target.hpp>
#include <migraph/gpu/target.hpp>
#include <migraph/gpu/miopen.hpp>
#include <migraph/gpu/hip.hpp>
...
...
@@ -100,7 +100,7 @@ migraph::argument run_cpu(migraph::program& p)
V
v
;
p
=
v
.
create_program
();
auto_print
pp
{
p
,
0
};
compile_check
(
p
,
migraph
::
cpu
::
cpu_
target
{});
compile_check
(
p
,
migraph
::
cpu
::
target
{});
migraph
::
program
::
parameter_map
m
;
for
(
auto
&&
x
:
p
.
get_parameter_shapes
())
{
...
...
Write
Preview
Markdown
is supported
0%
Try again
or
attach a new file
.
Attach a file
Cancel
You are about to add
0
people
to the discussion. Proceed with caution.
Finish editing this message first!
Cancel
Please
register
or
sign in
to comment