miopen.cpp 6.35 KB
Newer Older
Paul's avatar
Paul committed
1

Paul's avatar
Paul committed
2
3
4
5
#include <migraph/program.hpp>
#include <migraph/operators.hpp>
#include <migraph/generate.hpp>
#include <migraph/cpu/cpu_target.hpp>
Paul's avatar
Paul committed
6
7
8
#include <migraph/gpu/target.hpp>
#include <migraph/gpu/miopen.hpp>
#include <migraph/gpu/hip.hpp>
Paul's avatar
Paul committed
9
#include <migraph/manage_ptr.hpp>
Paul's avatar
Paul committed
10
11
12
13

#include <miopen/miopen.h>

#include "test.hpp"
Paul's avatar
Paul committed
14
#include "verify.hpp"
Paul's avatar
Paul committed
15

Paul's avatar
Paul committed
16
template <class V>
Paul's avatar
Paul committed
17
migraph::argument run_cpu()
Paul's avatar
Paul committed
18
{
Paul's avatar
Paul committed
19
20
    V v;
    auto p = v.create_program();
Paul's avatar
Paul committed
21
    p.compile(migraph::cpu::cpu_target{});
Paul's avatar
Paul committed
22
    migraph::program::parameter_map m;
Paul's avatar
Paul committed
23
    for(auto&& x : p.get_parameter_shapes())
Paul's avatar
Paul committed
24
25
26
27
    {
        m[x.first] = migraph::generate_argument(x.second);
    }
    return p.eval(m);
Paul's avatar
Paul committed
28
29
}

Paul's avatar
Paul committed
30
template <class V>
Paul's avatar
Paul committed
31
migraph::argument run_gpu()
Paul's avatar
Paul committed
32
{
Paul's avatar
Paul committed
33
34
    V v;
    auto p = v.create_program();
Paul's avatar
Paul committed
35
    p.compile(migraph::gpu::target{});
Paul's avatar
Paul committed
36

Paul's avatar
Paul committed
37
    migraph::program::parameter_map m;
Paul's avatar
Paul committed
38
    for(auto&& x : p.get_parameter_shapes())
Paul's avatar
Paul committed
39
    {
Paul's avatar
Paul committed
40
        m[x.first] = migraph::gpu::to_gpu(migraph::generate_argument(x.second));
Paul's avatar
Paul committed
41
42
    }

Paul's avatar
Paul committed
43
    return migraph::gpu::from_gpu(p.eval(m));
Paul's avatar
Paul committed
44
45
}

Paul's avatar
Paul committed
46
template <class V>
Paul's avatar
Paul committed
47
void verify_program()
Paul's avatar
Paul committed
48
{
Paul's avatar
Paul committed
49
50
    auto cpu_arg = run_cpu<V>();
    auto gpu_arg = run_gpu<V>();
Paul's avatar
Paul committed
51
    visit_all(cpu_arg, gpu_arg)([](auto cpu, auto gpu) { EXPECT(test::verify_range(cpu, gpu)); });
Paul's avatar
Paul committed
52
53
}

Paul's avatar
Paul committed
54
55
56
57
58
struct test_literals
{
    migraph::program create_program() const
    {
        migraph::program p;
Paul's avatar
Paul committed
59
60
61
62
        auto input = p.add_literal(
            generate_literal(migraph::shape{migraph::shape::float_type, {4, 3, 3, 3}}));
        auto weights = p.add_literal(
            generate_literal(migraph::shape{migraph::shape::float_type, {4, 3, 3, 3}}));
Paul's avatar
Paul committed
63
64
65
66
67
68
        auto conv = p.add_instruction(migraph::convolution{}, input, weights);
        p.add_instruction(migraph::activation{"relu"}, conv);
        return p;
    }
};

Paul's avatar
Paul committed
69
70
struct test_add
{
Paul's avatar
Paul committed
71
    migraph::program create_program() const
Paul's avatar
Paul committed
72
    {
Paul's avatar
Paul committed
73
74
        migraph::program p;
        migraph::shape s{migraph::shape::float_type, {3}};
Paul's avatar
Paul committed
75
76
        auto x = p.add_parameter("x", s);
        auto y = p.add_parameter("y", s);
Paul's avatar
Paul committed
77
        p.add_instruction(migraph::add{}, x, y);
Paul's avatar
Paul committed
78
79
80
81
82
83
        return p;
    }
};

struct test_add_broadcast
{
Paul's avatar
Paul committed
84
    migraph::program create_program() const
Paul's avatar
Paul committed
85
    {
Paul's avatar
Paul committed
86
87
88
89
90
91
        migraph::program p;
        migraph::shape s{migraph::shape::float_type, {3}};
        auto x  = p.add_parameter("x", {migraph::shape::float_type, {2, 2, 3}});
        auto y  = p.add_parameter("y", {migraph::shape::float_type, {2, 2}});
        auto by = p.add_instruction(migraph::broadcast{0}, x, y);
        p.add_instruction(migraph::add{}, x, by);
Paul's avatar
Paul committed
92
93
94
95
        return p;
    }
};

Paul's avatar
Paul committed
96
struct test_conv_relu
Paul's avatar
Paul committed
97
{
Paul's avatar
Paul committed
98
    migraph::program create_program() const
Paul's avatar
Paul committed
99
    {
Paul's avatar
Paul committed
100
        migraph::program p;
Paul's avatar
Paul committed
101
102
103
104
        auto input = p.add_parameter("x", migraph::shape{migraph::shape::float_type, {4, 3, 3, 3}});
        auto weights =
            p.add_parameter("w", migraph::shape{migraph::shape::float_type, {4, 3, 3, 3}});
        auto conv = p.add_instruction(migraph::convolution{}, input, weights);
Paul's avatar
Paul committed
105
        p.add_instruction(migraph::activation{"relu"}, conv);
Paul's avatar
Paul committed
106
107
108
109
        return p;
    }
};

Paul's avatar
Paul committed
110
111
struct test_conv_pooling
{
Paul's avatar
Paul committed
112
    migraph::program create_program() const
Paul's avatar
Paul committed
113
    {
Paul's avatar
Paul committed
114
        migraph::program p;
Paul's avatar
Paul committed
115
116
117
118
        auto input =
            p.add_parameter("x", migraph::shape{migraph::shape::float_type, {4, 3, 32, 32}});
        auto weights =
            p.add_parameter("w", migraph::shape{migraph::shape::float_type, {4, 3, 3, 3}});
Paul's avatar
Paul committed
119
120
121
        auto conv    = p.add_instruction(migraph::convolution{}, input, weights);
        auto pooling = p.add_instruction(migraph::pooling{"max"}, conv);
        p.add_instruction(migraph::activation{"relu"}, pooling);
Paul's avatar
Paul committed
122
123
124
125
        return p;
    }
};

Paul's avatar
Paul committed
126
127
struct test_gemm
{
Paul's avatar
Paul committed
128
    migraph::program create_program() const
Paul's avatar
Paul committed
129
    {
Paul's avatar
Paul committed
130
131
132
133
        migraph::program p;
        auto a = p.add_parameter("a", migraph::shape{migraph::shape::float_type, {4, 5}});
        auto b = p.add_parameter("b", migraph::shape{migraph::shape::float_type, {5, 3}});
        p.add_instruction(migraph::gemm{}, a, b);
Paul's avatar
Paul committed
134
135
136
137
        return p;
    }
};

138
139
140
141
142
143
144
145
146
147
148
149
150
151
152
153
154
155
156
157
158
159
160
161
162
163
164
165
166
167
168
169
170
171
172
173
174
175
176
177
struct test_gemm_transposeb
{
    migraph::program create_program() const
    {
        migraph::program p;
        auto a = p.add_parameter("a", migraph::shape{migraph::shape::float_type, {4, 5}});
        auto b = p.add_parameter("b", migraph::shape{migraph::shape::float_type, {3, 5}});
        auto bt = p.add_instruction(migraph::transpose{{1, 0}}, b);
        p.add_instruction(migraph::gemm{}, a, bt);
        return p;
    }
};

struct test_gemm_transposea
{
    migraph::program create_program() const
    {
        migraph::program p;
        auto a = p.add_parameter("a", migraph::shape{migraph::shape::float_type, {5, 4}});
        auto b = p.add_parameter("b", migraph::shape{migraph::shape::float_type, {5, 3}});
        auto at = p.add_instruction(migraph::transpose{{1, 0}}, a);
        p.add_instruction(migraph::gemm{}, at, b);
        return p;
    }
};

struct test_gemm_transposeab
{
    migraph::program create_program() const
    {
        migraph::program p;
        auto a = p.add_parameter("a", migraph::shape{migraph::shape::float_type, {5, 4}});
        auto b = p.add_parameter("b", migraph::shape{migraph::shape::float_type, {3, 5}});
        auto at = p.add_instruction(migraph::transpose{{1, 0}}, a);
        auto bt = p.add_instruction(migraph::transpose{{1, 0}}, b);
        p.add_instruction(migraph::gemm{}, at, bt);
        return p;
    }
};

178
179
180
181
182
struct test_contiguous
{
    migraph::program create_program() const
    {
        migraph::program p;
183
        migraph::shape s{migraph::shape::float_type, {4, 4, 4, 3}, {48, 4, 1, 16}};
184
185
186
187
188
189
        auto x = p.add_parameter("x", s);
        p.add_instruction(migraph::contiguous{}, x);
        return p;
    }
};

190
struct test_transpose
191
{
192
193
194
195
196
197
198
199
200
201
202
    migraph::program create_program() const
    {
        migraph::program p;
        migraph::shape s{migraph::shape::float_type, {4, 3, 4, 4}};
        auto x                    = p.add_parameter("x", s);
        std::vector<int64_t> perm = {0, 2, 3, 1};
        auto l                    = p.add_instruction(migraph::transpose{perm}, x);
        p.add_instruction(migraph::contiguous{}, l);
        return p;
    }
};
203

Paul's avatar
Paul committed
204
205
int main()
{
Paul's avatar
Paul committed
206
    verify_program<test_add>();
Paul's avatar
Paul committed
207
    verify_program<test_add_broadcast>();
Paul's avatar
Paul committed
208
209
210
    verify_program<test_conv_relu>();
    verify_program<test_conv_pooling>();
    verify_program<test_gemm>();
211
212
213
    verify_program<test_gemm_transposeb>();
    verify_program<test_gemm_transposea>();
    verify_program<test_gemm_transposeab>();
214
215
    verify_program<test_contiguous>();
    verify_program<test_transpose>();
Paul's avatar
Paul committed
216
}