doridoridoriand’s diary

主に技術的なことを書いていく予定(たぶん)

CPUで並列処理

今回はGPU使ってないのでこの画像は適切じゃないんだけど、ちょっと適当なの探すのが面倒だったのでｗ

C++でThread並列を実装してみました。といっても大層なことをしたわけではなく、OpenMPを利用しただけです。 そもそもOPenMPってなんぞやってかんじですよね。すごい雑にいうと、マルチスレッド対応の実行ファイルを生成するのを補助してくれる業界規格です。この辺の詳しい事に関しては、以下の資料を読むとわかりやすかったです

利用方法はとても簡単で、omp.hヘッダファイルをインクルードした上で、#pragma`ではじまるディレクティブをforやwhileの前に書いてあげるだけです

といってもコードを書かないと何言ってるんだってなると思われるので、まずシングルスレッド版(並列化していないバージョン)を示します

#include <iostream>
#include <string>
#include <stdlib.h>

const int REPEAT_TIMES = 10000;
unsigned int num_a[REPEAT_TIMES][REPEAT_TIMES], num_b[REPEAT_TIMES][REPEAT_TIMES], num_c[REPEAT_TIMES][REPEAT_TIMES];

void initialize_params() {
  for (int counter = 0; counter < REPEAT_TIMES; counter++) {
    for (int counter2 = 0; counter2 < REPEAT_TIMES; counter2++) {
      num_a[counter][counter2] = 1 * counter2;
      num_b[counter][counter2] = 2 * counter;
    }
  }
}

int calc_vector() {
  for (int counter = 0; counter < REPEAT_TIMES; counter++) {
    for (int counter2 = 0; counter2 < REPEAT_TIMES; counter2++) {
      num_c[counter][counter2] = num_a[counter][counter2] + num_b[counter][counter2];
    }
  }
}

int main (int argc, char** argv) {
  double measure_start, measure_stop;

  measure_start = clock();
  for (int counter = 0; counter <= 100; counter++) {
    initialize_params();
    calc_vector();
  }
  measure_stop = clock();

  std::cout << "処理時間" << (double)(measure_stop - measure_start) / CLOCKS_PER_SEC << "[s]" << std::endl;

  return 0;
}

まあ至って単純。ただの足し算をひたすらしているだけです。これを並列化するためには #include <omp.h>を新たに付け足し、適切な箇所でディレクティブを追記してあげればOK。 すると次のようになるかと思われます

#include <iostream>
#include <string>
#include <stdlib.h>
#include <omp.h>

//OpenMPを用いたCPU並列化バージョン

using namespace std;

const int REPEAT_TIMES = 10000;
unsigned int num_a[REPEAT_TIMES][REPEAT_TIMES], num_b[REPEAT_TIMES][REPEAT_TIMES], num_c[REPEAT_TIMES][REPEAT_TIMES];

void initialize_params() {
  #pragma omp parallel for
  for (int counter = 0; counter < REPEAT_TIMES; counter++) {
    #pragma omp parallel for
    for (int counter2 = 0; counter2 < REPEAT_TIMES; counter2++) {
      num_a[counter][counter2] = 1 * counter2;
      num_b[counter][counter2] = 2 * counter;
    }
  }
}

int calc_vector() {
  #pragma omp parallel for
  for (int counter = 0; counter < REPEAT_TIMES; counter++) {
    #pragma omp parallel for
    for (int counter2 = 0; counter2 < REPEAT_TIMES; counter2++) {
      num_c[counter][counter2] = num_a[counter][counter2] + num_b[counter][counter2];
    }
  }
}

int main (int argc, char** argv) {
  time_t start_time, stop_time;

  time(&start_time);
  #pragma omp parallel for
  for (int counter = 0; counter <= 100; counter++) {
    initialize_params();
    calc_vector();
  }
  time(&stop_time);

  cout << "処理時間" << (double)(stop_time - start_time) << "[s]" << endl;

  return 0;
}

今回はCentOS上で実行するので、コンパイルオプションに-fopenmpをつけてコンパイルしました

$ g++ ファイル名.cpp -o 出力したい名前 -fopenmp

これでコンパイルし、MacBookPro上に立てたVagrantで実行しました。
結果は以下の表の通りになりました

シングルスレッド: 53.9202[s]
マルチスレッド(4スレッド): 16.0[s]

おー。大体3.3倍くらい早くなっていますね。4倍とはいかないものの、効果は絶大です

すべての処理を並列化出来るわけではないですが(時間発展するやつとかはそこまで早くならない)、積極的にマルチスレッドを活用していきたいと思いました

PlayFrameworkをちょっと触った

PlayFrameworkを触る機会があったので、自分の備忘録がてら書きます 以下の操作はUbuntu上で実行しています

皆さんは知っていると思いますが、PlayFrameworkはScala(と一部Java)で実装されています。よって利用するにはJavaが必要になります。Ubuntuインストール時のオプションにも依りますが、Javaが入っていないこともあるので、まずはJavaをインストールします

PlayFrameworkの最新バージョンは、この記事を書いた時には2.4.2だったので、必要なJDKのバージョンは8でした。よってJDK8を入れます

$ sudo add-apt-repository ppa:webupd8team/java
$ sudo apt-get update
$ sudo apt-get install oracle-java8-installer
$ sudo apt-get install oracle-java8-set-default

次にPlayFramework本体をダウンロードしてきます。今回は本家のサイトからダウンロードしてくる形にしました

$ cd $HOME
$ wget "https://downloads.typesafe.com/typesafe-activator/1.3.5/typesafe-activator-1.3.5.zip"
$ unzip typesafe-activator-1.3.5.zip

実行ファイルに対してパスが通っていないので、環境変数に追記

export PATH=/home/ユーザー名/activator-dist-1.3.5/activator:$PATH

動作チェックを兼ねて、以下のコマンドで新規にPlayアプリケーションを作成します

$ activator new SampleApp

するとSampleAppというフォルダが出来たと思うの、移動して、以下のコマンドで実行します。初回起動はBuildが走るため、ちょっと時間がかかります

$ activator run

ブラウザからlocalhots:9000にアクセスして、Your new application is readyと表示されれば正常に稼働しています

雑記

いやー全然書いてねえなｗ
帰省している間はほとんどパソコンを開かなかったので、結局何も進みませんでしたとさ。まあ休みに行っているからそれでいいんだけど

OpenMPで並列化しようとしてうまく行かなかったりRubyの並列処理書き途中だったりして、色々と中途半端だったのですが、それよりも自分の部屋が最悪に汚くなって、完全に作業効率落ちていたので、今日はちょっと重めに掃除しました(大掃除ではないけど、いつもの掃除機かけるだけよりは上な感じで)

まずたくさん積まれた酒瓶を片付け、ビールの空き缶を片付け、ミネラルウォーターのペットボトルを捨てたらエライ綺麗になりました。つーかこまめに捨てよろって話ですね。すみません

さらにもう使わないであろう教科書やプリント類を捨てたり、ダンボールにまとめたら山みたいなのが消失しました。もっと早くにやればよかった。。。

僕の部屋はまだまだものが多いのですが、まあ少しずつ捨てるなり譲るなりしないとなと思っています。どうせ来年には引っ越すわけだし

というわけで久しぶりの更新でした

雑記

本当はRubyで並列計算に関してやろうと思ったのですが、Ruby標準のThreadクラスを使用して実装しようとしたところ思ったより手こずってしまったので、もうちょっと調査をしてから掲載します
なかなか面倒くさい。。。

僕は今帰省中なので普通に更新頻度が落ちます。更新しても雑記ばかりになることが予想されますので皆さんあまり期待しないでくださいｗ

まあだれも見てないかｗ

雑記

だれかNginxのリバースプロキシの設定教えてくれー。うごかんー

さて本当は色々書きたいことあるんだけど、記事にするにはまだ調査不足だし中途半端だしのネタを箇条書きにしてある(だけ)のエントリーだよ(ぉ

  • GIL(Global Intterspreter Lock)について色々な言語の対応状況まとめる
  • Cocos2dやりだしたし、せっかくならC++をもっとやりこもう。最近のC++すごいモダンだし(小並感)
  • HUGOのテーマ作る
  • サーバーサイドJavaScriptと真剣に向き合う
  • 無線LANルーター壊れかけたので、新しいの欲しい
  • PHPバージョン7ってなんぞや5.6(最新の安定版)より２倍早いらしいけど本当？？
  • CUDA ああCUDA
  • 広角の一眼のレンズ欲しい。何がいいかしら

・・・

それにしても暑いですね。僕は完全に夏バテしてしまいました。あと寝不足。あつい 個人的に静的サイトブームなので、ある程度までならCMSいらない感じのHUGOテーマつくろうって思っている(行動しているわけではない)

と言いつつ最近サーバーサイドで書いているのはSinatraですが。。。

GPUのまとめ3

いい加減コードを示していくよｗ

とりあえずコードを書きます 先にCUDAのドライバーなどがインストールされていること前提で進めています。そのうちインストール方法とかまとめます(ほんとか？？)

#include <iostream>
#include <malloc.h>
#include <helper_cuda.h>

# define N           (8192 * 8192)
# define block_size  196608
# define thread_size 512

__global__ void add( int *a, int *b, int *c) {
  int thread_id = blockIdx.x * blockDim.x + threadIdx.x;
  while (thread_id < N) {
    c[thread_id] = a[thread_id] + b[thread_id];
    thread_id += gridDim.x;
  }
}

int main(int argc, char** argv) {
  int *a, *b, *c;
  ind *dev_a, *dev_b, *dev_c;

  a = (int*)malloc(N * sizeof(int));
  b = (int*)malloc(N * sizeof(int));
  c = (int*)malloc(N * sizeof(int));

  cudaMalloc((void**)&dev_a, N * sizeof(int));
  cudaMalloc((void**)&dev_b, N * sizeof(int));
  cudaMalloc((void**)&dev_c, N * sizeof(int));

  for (int i = 0; i < N; i++) {
    a[i] = i;
    b[i] = 2 * i;
  }

  cudaMemcpy(dev_a, a, N * siezeof(int), cudaMemcpyHostToDevice);
  cudaMemcpy(dev_b, b, N * siezeof(int), cudaMemcpyHostToDevice);

  add<<<block_size, thread_size>>>(dev_a, deb_b, dev_c);

  cudamemcpy(c, dev_c, N * sizeof(int), cudaMemcpyDeviceToHost);

  cudaFree(dev_a);
  cudaFree(dev_b);
  cudaFree(dev_c);

  free(a);
  free(b);
  free(c);

  return 0;
}

上から説明していきます。CUDA独自の書き方とかまあ色々説明しなければならないことはあるのですが、まあ適宜最小限の説明で行きます。 最初に色々言われても実際のコード見ないとわからないこと多いもんね

# define N           (8192 * 8192)
# define block_size  196608
# define thread_size 512

まずここの定数宣言なんだって感じですよね

# define N (8192 * 8192)

今回はCUDAの性能を純粋に測りたかったので単純なベクトル和を計算するのを書いています。 よってこれは計算するベクトルの総和。つまり計算対象は 8192 * 8192 あるってことです

# define block_size  196608

使用するブロック数です。これテキトー(というか、さっきまで使ってたGPUに合わせただけ)なので、適宜変えて大丈夫です。これを変えて処理時間とか見みると結構面白い

define thread_size 512

使用するスレッド数です。これも変えておｋ

__global__ void add( int *a, int *b, int *c) {
  int thread_id = blockIdx.x * blockDim.x + threadIdx.x;
  while (thread_id < N) {
    c[thread_id] = a[thread_id] + b[thread_id];
    thread_id += gridDim.x;
  }
}

これが今回のメインの話題となるやつですね

__global__ってついたやつはコンパイル時にGPU用のコードとしてコンパイルされます。 もっと正確にいうと、コンパイルはnvccというnVIDIAのコンパイラを使うのですが、コンパイルの流れとしては

  • コンパイル実行 nvcc -o hoge hoge.cu
  • nvccがCPUコードとGPUコードを判別。CPU用はgccに渡し、GPU用はnvccが担当する
  • 最後CPU側がGPUバイトコードを呼ぶような感じでリンクして実行ファイル生成(めっちゃ適当)

ってなっていますGPU側はPTXコードというのが生成されたり色々するのですが、 僕があんま詳しくないのでもう少し詳しく取り扱いたいのでまた今度

int thread_id = blockIdx.x * blockDim.x + threadIdx.x;

これがGPUの特徴をよく表していると個人的に思うのですが、並列計算をするときに、SP同士でメモリの参照が狂ったりしないようにそれぞれの順番をuniqになるように定義できるのです。 ちなみに使用するスレッドが１つだけの場合、つまり# define thread_size 1のときはthreadIdxはいらなくて良かったりします

以下はとくに変ではない、ふつうのCに登場してくるwhile文なので説明は省略します。
ではメイン関数をば

int *a, *b, *c;
ind *dev_a, *dev_b, *dev_c;

初期化はポインタで指定してね。でないと参照渡しとか出来ないので

cudaMalloc((void**)&dev_a, N * sizeof(int));
cudaMalloc((void**)&dev_b, N * sizeof(int));
cudaMalloc((void**)&dev_c, N * sizeof(int));

これでcuda側のメモリ確保を行います。

for (int i = 0; i < N; i++) {
  a[i] = i;
  b[i] = 2 * i;
}

ベクトル和の値の初期化をしてるだけです

cudaMemcpy(dev_a, a, N * siezeof(int), cudaMemcpyHostToDevice);

これでCPU側で初期化した配列群をGPU側のグローバルメモリにコピーします

add<<<block_size, thread_size>>>(dev_a, deb_b, dev_c);

ここで先ほどのGPUのコードを呼び出しています。 <<<>>>で囲まれた部分でこの関数で使用するブロックとスレッドの数を指定できます

cudamemcpy(c, dev_c, N * sizeof(int), cudaMemcpyDeviceToHost);

ここで計算し終わったデータをGPU側からCPU側にコピーしています

cudaFree(dev_a);

みんなだいすきメモリ解放。Cのfree()のCUDA版です

すげーどんどん説明がテキトーになっているのがわかるかと思われますが、夏バテなので許してください

最後に__global__の関数、つまりCUDAで処理される関数はvoid型しか取れないのでみなさん注意してください

次回はAWSで実行環境を作る話をしたいです。僕はもうお家帰って寝たいです