miopen.cpp 6.49 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
#include <migraph/type_name.hpp>
Paul's avatar
Paul committed
11
12
13
14

#include <miopen/miopen.h>

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

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

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

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

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

Paul's avatar
Paul committed
47
template <class V>
Paul's avatar
Paul committed
48
void verify_program()
Paul's avatar
Paul committed
49
{
Paul's avatar
Paul committed
50
51
    auto cpu_arg = run_cpu<V>();
    auto gpu_arg = run_gpu<V>();
Paul's avatar
Paul committed
52
53
54
55
56
    visit_all(cpu_arg, gpu_arg)([](auto cpu, auto gpu) { 
        if(not test::verify_range(cpu, gpu)) {
            std::cout << "FAILED: " << migraph::get_type_name<V>() << std::endl;
        }
    });
Paul's avatar
Paul committed
57
58
}

Paul's avatar
Paul committed
59
60
61
62
63
struct test_literals
{
    migraph::program create_program() const
    {
        migraph::program p;
Paul's avatar
Paul committed
64
65
66
67
        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
68
69
70
71
72
73
        auto conv = p.add_instruction(migraph::convolution{}, input, weights);
        p.add_instruction(migraph::activation{"relu"}, conv);
        return p;
    }
};

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

struct test_add_broadcast
{
Paul's avatar
Paul committed
89
    migraph::program create_program() const
Paul's avatar
Paul committed
90
    {
Paul's avatar
Paul committed
91
92
93
94
95
96
        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
97
98
99
100
        return p;
    }
};

Paul's avatar
Paul committed
101
struct test_conv_relu
Paul's avatar
Paul committed
102
{
Paul's avatar
Paul committed
103
    migraph::program create_program() const
Paul's avatar
Paul committed
104
    {
Paul's avatar
Paul committed
105
        migraph::program p;
Paul's avatar
Paul committed
106
107
108
109
        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
110
        p.add_instruction(migraph::activation{"relu"}, conv);
Paul's avatar
Paul committed
111
112
113
114
        return p;
    }
};

Paul's avatar
Paul committed
115
116
struct test_conv_pooling
{
Paul's avatar
Paul committed
117
    migraph::program create_program() const
Paul's avatar
Paul committed
118
    {
Paul's avatar
Paul committed
119
        migraph::program p;
Paul's avatar
Paul committed
120
121
122
123
        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
124
125
126
        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
127
128
129
130
        return p;
    }
};

Paul's avatar
Paul committed
131
132
struct test_gemm
{
Paul's avatar
Paul committed
133
    migraph::program create_program() const
Paul's avatar
Paul committed
134
    {
Paul's avatar
Paul committed
135
136
137
138
        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
139
140
141
142
        return p;
    }
};

143
144
145
146
147
struct test_gemm_transposeb
{
    migraph::program create_program() const
    {
        migraph::program p;
Paul's avatar
Paul committed
148
149
        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}});
150
151
152
153
154
155
156
157
158
159
160
        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;
Paul's avatar
Paul committed
161
162
        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}});
163
164
165
166
167
168
169
170
171
172
173
        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;
Paul's avatar
Paul committed
174
175
        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}});
176
177
178
179
180
181
182
        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;
    }
};

183
184
185
186
187
struct test_contiguous
{
    migraph::program create_program() const
    {
        migraph::program p;
188
        migraph::shape s{migraph::shape::float_type, {4, 4, 4, 3}, {48, 4, 1, 16}};
189
190
191
192
193
194
        auto x = p.add_parameter("x", s);
        p.add_instruction(migraph::contiguous{}, x);
        return p;
    }
};

195
struct test_transpose
196
{
197
198
199
200
201
202
203
204
205
206
207
    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;
    }
};
208

Paul's avatar
Paul committed
209
210
int main()
{
Paul's avatar
Paul committed
211
    verify_program<test_add>();
Paul's avatar
Paul committed
212
    verify_program<test_add_broadcast>();
Paul's avatar
Paul committed
213
214
215
    verify_program<test_conv_relu>();
    verify_program<test_conv_pooling>();
    verify_program<test_gemm>();
216
217
218
    verify_program<test_gemm_transposeb>();
    verify_program<test_gemm_transposea>();
    verify_program<test_gemm_transposeab>();
219
220
    verify_program<test_contiguous>();
    verify_program<test_transpose>();
Paul's avatar
Paul committed
221
}