13
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
1 <!DOCTYPE html>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
2
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
3 <html>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
4 <head>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
5 <title>Presentation</title>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
6
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
7 <meta charset='utf-8'>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
8 <script
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
9 src='./slides.js'></script>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
10 </head>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
11
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
12 <style>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
13 /* Your individual styles here, or just use inline styles if that’s
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
14 what you want. */
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
15
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
16 </style>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
17
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
18 <body style='display: none'>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
19 <section class='slides layout-regular template-default'>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
20
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
21 <!--
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
22 Your slides (<article>s) go here. Delete or comment out the
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
23 slides below.
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
24 -->
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
25 <article >
|
14
|
26 <h1>Cerium による並列処理向け I/O の設計と実装</h1>
|
|
27 <h3 class="title">Masataka Kohagura 12th, February</h3>
|
13
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
28 <div align="right">担当教官 : 河野 真治</div>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
29 </article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
30 <article >
|
14
|
31 <h3>研究背景と目的</h3>
|
13
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
32 <p>
|
14
|
33 近年のCPUのほとんどはマルチコアであり、それらの性能を引き出すためには並列プログラミングが必須となっている。そこで当研究室では Cerium Library の開発を行い、提供することによって並列プログラミングを容易にしている。
|
13
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
34 </p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
35 <p>
|
14
|
36 先行研究では Task の並列化によって、プログラム全体の処理速度は向上しているが、ファイル読み込み等の I/O に対して並列に Task が動作するようにはされていない。
|
13
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
37 </p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
38 <p>
|
14
|
39 本研究では I/O と Task の並列化の設計、実装によってプログラム全体の処理速度、処理効率を上げていく。
|
13
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
40 </p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
41 </article>
|
14
|
42
|
13
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
43 <article>
|
14
|
44 <h3>Cerium とは</h3>
|
13
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
45 <table border="0" cellpadding="0" cellspacing="0">
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
46 <tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
47 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
48 <td><img src='images/createtask.png' style="height:350px"></td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
49 <td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
50 <ol>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
51 <li>Taskを生成</li>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
52 <li>依存関係のチェック</li>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
53 <li>Schedulerに転送</li>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
54 <li>並列実行</li>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
55 </ol>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
56 </td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
57 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
58 </tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
59 </table>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
60 <p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
61 CpuThreads、Schedulerに対応させる形でGpuThreadsとGpuSchedulerを作成した
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
62 </p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
63 </article>
|
14
|
64
|
|
65
|
13
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
66 <article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
67 <h3>GPU Task実行の流れ</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
68 <br>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
69 <h3 class="yellow">kernel fileの記述</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
70 <pre>__kernel void // kernel.cl(kernel file)
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
71 twice(__global int *input_data,__global int *output_data) {
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
72 long count = (long)data_count[0];
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
73 for (int i = 0; i< count; i++) {
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
74 output_data[i] = input_data[i] * 2;;
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
75 }
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
76 }
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
77 </pre>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
78 </article>
|
14
|
79
|
13
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
80 <article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
81 <h3>GPU Task 実行の流れ</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
82 <br>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
83 <h3 class="yellow">kernelをTaskとしてCeriumに登録</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
84 <pre>void
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
85 task_init(void) { // task_init.cc
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
86 GpuSchedRegister(Twice, "./twice.cl", "twice");
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
87 }</pre>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
88 <table border="2" style="font-size:18pt;">
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
89 <tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
90 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
91 <td> 第1引数<br>Twice</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
92 <td >Taskのid。enumで定義されている<br>TaskManagerはこの値でTaskを識別する</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
93 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
94 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
95 <td> 第2引数<br>"./twice.cl"</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
96 <td>OpenCLが処理するkernelが記述されているファイルのパス</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
97 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
98 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
99 <td> 第3引数<br>"twice"</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
100 <td >関数の指定。kernel file内にある、実行する関数名を指定<br>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
101 Taskにあたる部分</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
102 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
103 </tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
104 </table>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
105 </article>
|
14
|
106
|
13
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
107 <article class="nobackground">
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
108 <h3>GPU Task 実行の流れ</h3><br>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
109 <h3 class="yellow">GPU Task生成</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
110 <pre>// main.cc
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
111 HTaskPtr twice = manager->create_task(Twice);
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
112 twice->set_inData(0, indata, sizeof (int)*length);
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
113 // twice->set_inData(1, indata2, sizeof (int)*length);
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
114 twice->set_outData(0, outdata, sizeof (int)*length);
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
115 twice->set_cpu(GPU);
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
116 twice->spawn();
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
117 </pre>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
118 </article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
119
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
120 <article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
121 <h3>Cerium OpenCL API比較</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
122 <img src='images/api.png' style="height:500px">
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
123 </article>
|
14
|
124
|
13
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
125 <article class="nobackground">
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
126 <h3>ベンチマーク</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
127 <table >
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
128 <tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
129 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
130 <td> <!-- benchmark -->
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
131 <table border="2" style="font-size:18pt;">
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
132 <tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
133 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
134 <td bgcolor="#8091B5"></td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
135 <td style="text-align: center;">Time</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
136 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
137 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
138 <td style="text-align: center;">1 CPU</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
139 <td style="text-align: right;"> 796 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
140 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
141 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
142 <td style="text-align: center;">2 CPU</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
143 <td style="text-align: right;"> 439 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
144 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
145 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
146 <td style="text-align: center;">6 CPU</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
147 <td style="text-align: right;"> 153 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
148 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
149 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
150 <td style="text-align: center;">12 CPU</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
151 <td style="text-align: right;"> 96 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
152 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
153 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
154 <td style="text-align: center;">24 CPU</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
155 <td style="text-align: right;"> 89 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
156 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
157 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
158 <td style="text-align: center;" bgcolor="#ffffcc">GPU(改良前)</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
159 <td style="text-align: right; " bgcolor="#ffffcc"> 330752 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
160 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
161 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
162 <td style="text-align: center;" bgcolor="#ffffcc">GPU(改良後)</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
163 <td style="text-align: right;" bgcolor="#ffffcc"> 5306 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
164 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
165 </tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
166 </table>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
167 </td> <!-- /benchmark -->
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
168 <td> <!-- system env -->
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
169 <h3 class="yellow">10万入力によるBitonic Sort</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
170 <font size="5">
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
171 <p>実験環境</p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
172 OS : MacOS 10.8.2<br>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
173 CPU : 2*2.66GHz 6-CoreIntel Xeon<br>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
174 Memory : 16GB<br>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
175 Compiler : Apple clang version 4.1 <br> (based on LLVM 3.1svn)<br>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
176 GPU : AMD ATI Radeon HD 5870 1024MB<br>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
177 </font>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
178 <h3 class="yellow">結果</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
179 1coreのCPUよりも10倍遅い
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
180 </td> <!--system env -->
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
181 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
182 </tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
183 </table>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
184 <p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
185 充分な性能が出なかったため、一度に送信する
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
186 data のサイズを増やす改善を行ったところ、
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
187 <font color="red">約60倍</font>程実行速度が向上した
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
188 </p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
189 </article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
190 <article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
191 <h3>考察</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
192 <p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
193 性能向上は見られたが、CPUと比べると未だ差が開いている
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
194 GPU向けに適切なチューニングが今後の課題となる
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
195 </p><br>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
196 <h3 class="yellow">改善案</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
197 <ul>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
198 <li>データ並列によるkernelの実行</li>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
199 <li>同期機構の見直し</li>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
200 </ul>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
201 </article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
202 <article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
203 <h3>データ並列</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
204 <p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
205 データを2、3次元に分割し、分割した部分に対して並列処理する並列化手法。
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
206 </p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
207 <p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
208 OpenCL ではin/outするデータ郡をWork Itemと呼ぶ。
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
209 </p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
210 <table border="0" cellpadding="0" cellspacing="0">
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
211 <tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
212 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
213 <td><img src='images/ndrange_arch.png' style="height:350px"></td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
214 <td>各Work Item のサイズを指定するとOpenCLがデータ並列で実行する。</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
215 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
216 </tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
217 </table>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
218 </article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
219 <article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
220 <h3>同期機構</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
221 <p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
222 GpuSchedulerはCommand Queueの内部でパイプライン的に実行を行っている。
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
223 パイプラインを構成するには処理にwaitをかける必要がある。現在はclWaitForEvent APIを使用
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
224 </p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
225 <table border="0" cellpadding="0" cellspacing="0">
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
226 <tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
227 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
228 <td bgcolor="#8091B5"><font color="white">API</font></td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
229 <td>機能</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
230 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
231 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
232 <td bgcolor="#8091B5"><font color="white">clFlush()</font></td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
233 <td>Command Queueに投入したTask全てをDeviceで実行する</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
234 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
235 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
236 <td bgcolor="#8091B5"><font color="white">clWaitForEvent()</font></td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
237 <td>特定の処理の終了を待つ</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
238 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
239 </tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
240 </table>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
241 <p>clFlushは実行は保証するが、<font color="red">終了は保証しない</font>仕様になっている</p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
242 </article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
243 <article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
244 <h3>新しい同期</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
245 <table border="0" cellpadding="0" cellspacing="0">
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
246 <tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
247 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
248 <td bgcolor="#8091B5"><font color="white">FrameWork</font></td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
249 <td>Dependency</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
250 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
251 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
252 <td bgcolor="#8091B5"><font color="white">Cerium</font></td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
253 <td>Task Dependency</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
254 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
255 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
256 <td bgcolor="#8091B5"><font color="white">OpenCL</font></td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
257 <td>Data Dependency</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
258 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
259 </tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
260 </table>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
261 <p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
262 Task Dependency:Schedulerで依存関係が決定<br>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
263 Data Dependency:GPUに読み込まれた時に決定
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
264 </p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
265 <p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
266 GPGPUはなるべくGPU内部で処理を行う方が高速なため、性能向上が見込める
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
267 </p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
268 </article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
269 <article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
270 <h3>まとめ</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
271 <ul>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
272 <li>Cerium Task ManagerをGPGPUに対応</li>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
273 <li>同期機構の実装</li>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
274 <li>マルチコア実行とGPU実行のベンチマーク</li>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
275 </ul>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
276 <h3 class="yellow">今後の課題</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
277 <ul>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
278 <li>データ並列による実行のサポート</li>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
279 <li>同期機構の見直し</li>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
280 </ul>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
281 </article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
282 <article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
283 <h3>ベンチマーク</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
284 <table >
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
285 <tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
286 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
287 <td> <!-- benchmark -->
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
288 <table border="2" style="font-size:18pt;">
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
289 <tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
290 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
291 <td bgcolor="#8091B5"></td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
292 <td style="text-align: center;">Time</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
293 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
294 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
295 <td style="text-align: center;">1 CPU</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
296 <td style="text-align: right;"> 67 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
297 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
298 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
299 <td style="text-align: center;">2 CPU</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
300 <td style="text-align: right;"> 34 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
301 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
302 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
303 <td style="text-align: center;">6 CPU</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
304 <td style="text-align: right;"> 12 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
305 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
306 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
307 <td style="text-align: center;">12 CPU</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
308 <td style="text-align: right;"> 9 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
309 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
310 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
311 <td style="text-align: center;">24 CPU</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
312 <td style="text-align: right;"> 6 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
313 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
314 <tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
315 <td style="text-align: center;" bgcolor="#ffffcc">GPU</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
316 <td style="text-align: right;" bgcolor="#ffffcc"> 10201 ms</td>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
317 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
318 </tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
319 </table>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
320 </td> <!-- /benchmark -->
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
321 <td> <!-- system env -->
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
322 <h3 class="yellow">word count</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
323 <font size="5">
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
324 <p>10MBのテキストファイルを分割<br>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
325 各Taskがcountしていく</p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
326 <p>スペースと改行区切りでword countしていく</p>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
327 </font>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
328 <h3 class="yellow">結果</h3>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
329 CPUの方が150倍早い
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
330 </td> <!--system env -->
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
331 </tr>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
332 </tbody>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
333 </table>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
334 </article>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
335 <!--- <img src='images/flow_chart.jpg' width="300" height="500"> -->
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
336 </body>
|
Masataka Kohagura <e085726@ie.u-ryukyu.ac.jp>
parents:
diff
changeset
|
337 </html>
|