view presen/index.html @ 8:954053cacd0e default tip

add resume
author Yuhi TOMARI <yuhi@cr.ie.u-ryukyu.ac.jp>
date Thu, 21 Feb 2013 19:29:57 +0900
parents 84766a598c3a
children
line wrap: on
line source

<!DOCTYPE html>

<html>
  <head>
    <title>Presentation</title>
    
    <meta charset='utf-8'>
    <script
       src='./slides.js'></script>
  </head>

  <style>
    /* Your individual styles here, or just use inline styles if that’s
    what you want. */

  </style>

  <body style='display: none'>
    <section class='slides layout-regular template-default'>

      <!-- 
           Your slides (<article>s) go here. Delete or comment out the
           slides below.
        -->
      <article >
        <h1>Cerium TaskManagerのGPGPUのサポート</h1>
        <h3 class="title">Yuhi TOMARI  21 Feb 2013</h3>
        <div align="right">担当教官 : 河野 真治</div>
      </article>
      <article >
        <h3>研究概要</h3>
        <p>
          当研究室ではCellおよびLinux、
          Mac OSX上で動く並列プログラミングフレームワーク、
          Ceriumの開発・改良を行っている
        </p>
        <img src='images/cerium_resource.png' style="height:300px">
        <p>
          <font color="#ff3300">本研究では新たにGPU上での並列実行に対応</font>
        </p>
        <p>
          これによりヘテロジニアス(異種混合)環境下<br>でのプログラミングをサポートする
        </p>

        <!--
            <p>
              CeriumにOpenCLを組み込む事でGpu上での並列実行にも対応すると、
              CPU・GPUそれぞれにtaskを割り振ることが可能となる。</p>
        <p>
          これにより、ヘテロジニアス(異種混合)な環境で
          それぞれのアーキテクチャに適したプログラミングをサポートする。
        </p>
        -->
        
      </article>
      <article>
        <h3>OpenCLによるGPGPUへの対応</h3>
        <p>
          OpenCLは
          ヘテロジニアス環境での並列計算を支援するフレームワーク
        </p>
        <table  border="0" cellpadding="0" cellspacing="0">
          <tbody>
            <tr>
              <td><img src='images/opencl.png' style="height:120px"></td>
              <td>
                <p>
                  OpenCLのtaskはkernelと呼ばれ、OpenCL Cという
                  Cに似た独自の言語で記述される
                </p>
              </td>
            </tr>
          </tbody>
        </table>
        <p>          
          <h3 class='yellow'>CPU/Cell上で実行する場合とGPU上で実行する場合、<br>
            それぞれほぼ同じ形式でkernelを記述できる</h3>
        </p>
      </article>
      <article>
        <h3>Cerium Task Manager</h3>
        <table  border="0" cellpadding="0" cellspacing="0">
          <tbody>
            <tr>
              <td><img src='images/createtask.png' style="height:350px"></td>
              <td>
                <ol>
                  <li>Taskを生成</li>
                  <li>依存関係のチェック</li>
                  <li>Schedulerに転送</li>
                  <li>並列実行</li>
                </ol>
              </td>
            </tr>
          </tbody>
        </table>
        <p>
          CpuThreads、Schedulerに対応させる形でGpuThreadsとGpuSchedulerを作成した
        </p>
      </article>
      <article class="nobackground">
        <h3>GPUのアーキテクチャ</h3>
          <table  border="0" cellpadding="0" cellspacing="0">
            <tbody>
              <tr>
                <td><img src='images/gpu_arch.png' style="height:400px"></td>
                <td>
                  <p>
                    GPU上のローカルメモリはCPUとメモリ空間が異なるのでメモリの共有ができない。
                  </p>
                </td>
              </tr>
            </tbody>
          </table>
          <p>
            メモリのコピー処理やTaskを順にCommand Queueに
            enqueueする。enqueueされた処理は並列実行される
      </article>
      <article>
        <h3>GPU Task実行の流れ</h3>
          <br>
          <h3 class="yellow">kernel fileの記述</h3>
<pre>__kernel void // kernel.cl(kernel file)
twice(__global int *input_data,__global int *output_data) {
    long count = (long)data_count[0];
    for (int i = 0; i&lt count; i++) {
        output_data[i] = input_data[i] * 2;;
    }
}
</pre>
      </article>
      <article>
        <h3>GPU Task 実行の流れ</h3>
        <br>
        <h3 class="yellow">kernelをTaskとしてCeriumに登録</h3>
        <pre>void
task_init(void) { // task_init.cc
    GpuSchedRegister(Twice, "./twice.cl", "twice");
}</pre>
        <table  border="2" style="font-size:18pt;">
          <tbody>
            <tr>
              <td> 第1引数<br>Twice</td>
              <td >Taskのid。enumで定義されている<br>TaskManagerはこの値でTaskを識別する</td>
            </tr>
            <tr>
              <td> 第2引数<br>"./twice.cl"</td>
              <td>OpenCLが処理するkernelが記述されているファイルのパス</td>
            </tr>
            <tr>
              <td> 第3引数<br>"twice"</td>
              <td >関数の指定。kernel file内にある、実行する関数名を指定<br>
                Taskにあたる部分</td>
            </tr>
          </tbody>
        </table>
      </article>
      <article class="nobackground">
        <h3>GPU Task 実行の流れ</h3><br>
        <h3 class="yellow">GPU Task生成</h3>
<pre>// main.cc
HTaskPtr twice = manager->create_task(Twice);
twice->set_inData(0, indata, sizeof (int)*length);
// twice->set_inData(1, indata2, sizeof (int)*length);
twice->set_outData(0, outdata, sizeof (int)*length);
twice->set_cpu(GPU);
twice->spawn();
        </pre>
      </article>

      
      <article>
        <h3>Cerium OpenCL API比較</h3>
      <img src='images/api.png' style="height:500px">  
      </article>
      <article class="nobackground">
        <h3>ベンチマーク</h3>
        <table >
          <tbody>
            <tr>
              <td> <!--  benchmark -->
                <table  border="2" style="font-size:18pt;">
                  <tbody>
                    <tr>
                      <td bgcolor="#8091B5"></td>
                      <td style="text-align: center;">Time</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;">1 CPU</td>
                      <td style="text-align: right;"> 796 ms</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;">2 CPU</td>
                      <td style="text-align: right;"> 439 ms</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;">6 CPU</td>
                      <td style="text-align: right;"> 153 ms</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;">12 CPU</td>
                      <td style="text-align: right;"> 96 ms</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;">24 CPU</td>
                      <td style="text-align: right;"> 89 ms</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;" bgcolor="#ffffcc">GPU(改良前)</td>
                      <td style="text-align: right; " bgcolor="#ffffcc"> 330752 ms</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;" bgcolor="#ffffcc">GPU(改良後)</td>
                      <td style="text-align: right;" bgcolor="#ffffcc"> 5306 ms</td>
                    </tr>
                  </tbody>
                </table>
              </td> <!-- /benchmark -->
              <td>  <!-- system env -->
                <h3 class="yellow">10万入力によるBitonic Sort</h3>
                <font size="5">
                  <p>実験環境</p>
                  OS : MacOS 10.8.2<br>
                  CPU : 2*2.66GHz 6-CoreIntel Xeon<br>
                  Memory : 16GB<br>
                  Compiler : Apple clang version 4.1 <br>     (based on LLVM 3.1svn)<br>
                  GPU : AMD ATI Radeon HD 5870 1024MB<br>
                </font>
                <h3 class="yellow">結果</h3>
                1coreのCPUよりも10倍遅い
              </td> <!--system env  -->
            </tr>
          </tbody>
        </table>
        <p>
          充分な性能が出なかったため、一度に送信する
          data のサイズを増やす改善を行ったところ、
          <font color="red">約60倍</font>程実行速度が向上した
        </p>
      </article>
      <article>
        <h3>考察</h3>
        <p>
          性能向上は見られたが、CPUと比べると未だ差が開いている
          GPU向けに適切なチューニングが今後の課題となる
        </p><br>
        <h3 class="yellow">改善案</h3>
        <ul>
          <li>データ並列によるkernelの実行</li>
          <li>同期機構の見直し</li>
        </ul>
      </article>
      <article>
        <h3>データ並列</h3>
        <p>
          データを2、3次元に分割し、分割した部分に対して並列処理する並列化手法。
        </p>
        <p>
          OpenCL ではin/outするデータ郡をWork Itemと呼ぶ。
        </p>
        <table  border="0" cellpadding="0" cellspacing="0">
          <tbody>
            <tr>
              <td><img src='images/ndrange_arch.png' style="height:350px"></td>
              <td>各Work Item のサイズを指定するとOpenCLがデータ並列で実行する。</td>
            </tr>
          </tbody>
        </table>
      </article>
      <article>
        <h3>同期機構</h3>
        <p>
          GpuSchedulerはCommand Queueの内部でパイプライン的に実行を行っている。
          パイプラインを構成するには処理にwaitをかける必要がある。現在はclWaitForEvent APIを使用
        </p>
        <table  border="0" cellpadding="0" cellspacing="0">
          <tbody>
            <tr>
              <td bgcolor="#8091B5"><font color="white">API</font></td>
              <td>機能</td>
            </tr>
            <tr>
              <td bgcolor="#8091B5"><font color="white">clFlush()</font></td>
              <td>Command Queueに投入したTask全てをDeviceで実行する</td>
            </tr>
            <tr>
              <td bgcolor="#8091B5"><font color="white">clWaitForEvent()</font></td>
              <td>特定の処理の終了を待つ</td>
            </tr>
          </tbody>
        </table>
        <p>clFlushは実行は保証するが、<font color="red">終了は保証しない</font>仕様になっている</p>
      </article>
      <article>
        <h3>新しい同期</h3>
        <table  border="0" cellpadding="0" cellspacing="0">
          <tbody>
            <tr>
              <td bgcolor="#8091B5"><font color="white">FrameWork</font></td>
              <td>Dependency</td>
            </tr>
            <tr>
              <td bgcolor="#8091B5"><font color="white">Cerium</font></td>
              <td>Task Dependency</td>
            </tr>
            <tr>
              <td bgcolor="#8091B5"><font color="white">OpenCL</font></td>
              <td>Data Dependency</td>
            </tr>
          </tbody>
        </table>
        <p>
          Task Dependency:Schedulerで依存関係が決定<br>
          Data Dependency:GPUに読み込まれた時に決定
        </p>
        <p>
          GPGPUはなるべくGPU内部で処理を行う方が高速なため、性能向上が見込める
        </p>
      </article>
      <article>
        <h3>まとめ</h3>
        <ul>
          <li>Cerium Task ManagerをGPGPUに対応</li>
          <li>同期機構の実装</li>
          <li>マルチコア実行とGPU実行のベンチマーク</li>
        </ul>
        <h3 class="yellow">今後の課題</h3>
        <ul>
          <li>データ並列による実行のサポート</li>
          <li>同期機構の見直し</li>
        </ul>
      </article>
      <article>
        <h3>ベンチマーク</h3>
        <table >
          <tbody>
            <tr>
              <td> <!--  benchmark -->
                <table border="2" style="font-size:18pt;">
                  <tbody>
                    <tr>
                      <td bgcolor="#8091B5"></td>
                      <td style="text-align: center;">Time</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;">1 CPU</td>
                      <td style="text-align: right;"> 67 ms</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;">2 CPU</td>
                      <td style="text-align: right;"> 34 ms</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;">6 CPU</td>
                      <td style="text-align: right;"> 12 ms</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;">12 CPU</td>
                      <td style="text-align: right;"> 9 ms</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;">24 CPU</td>
                      <td style="text-align: right;"> 6 ms</td>
                    </tr>
                    <tr>
                      <td style="text-align: center;" bgcolor="#ffffcc">GPU</td>
                      <td style="text-align: right;" bgcolor="#ffffcc"> 10201 ms</td>
                    </tr>
                  </tbody>
                </table>
              </td> <!-- /benchmark -->
              <td>  <!-- system env -->
                <h3 class="yellow">word count</h3>
                <font size="5">
                  <p>10MBのテキストファイルを分割<br>
                    各Taskがcountしていく</p>
                  <p>スペースと改行区切りでword countしていく</p>
                </font>
                <h3 class="yellow">結果</h3>
                CPUの方が150倍早い
              </td> <!--system env  -->
            </tr>
          </tbody>
        </table>
      </article>
      <!--- <img src='images/flow_chart.jpg' width="300" height="500">  -->
  </body>
</html>