GPGPU#5

**デフォルトの名無しさん** · 2010/08/15(日) 21:47:50

GPGPUについて語りましょう

前スレ
GPGPU#4
http://hibari.2ch.net/test/read.cgi/tech/1255256230/l50

関連スレ
OpenCLプログラミング#1
http://hibari.2ch.net/test/read.cgi/tech/1228891105/l50
【GPGPU】くだすれCUDAスレ pert3【NVIDIA】
http://hibari.2ch.net/test/read.cgi/tech/1271587710/l50

参考リンク
総本山? gpgpu.org
http://www.gpgpu.org/
OpenCL
http://www.khronos.org/opencl/
NVIDIA CUDA
http://developer.nvidia.com/object/cuda.html
ATI Stream
http://developer.amd.com/gpu/ATIStreamSDK/Pages/default.aspx
GPUをCPU的に活用するGPGPUの可能性
http://pcweb.mycom.co.jp/articles/2005/09/06/siggraph2/

**デフォルトの名無しさん** · 2012/10/14(日) 11:35:52.75

>>241
亀レスだけど、何のペナルティも無く普通のx86コアを50個
並べられるなら誰も苦労してないわけで･･･

**デフォルトの名無しさん** · 2012/10/14(日) 12:44:22.63

DRAM内部のクロック
PC133　133MHz
DDR3-1600　200MHz
この間十数年・・・MRAMならもっと上がるかな？

**デフォルトの名無しさん** · 2012/10/20(土) 09:38:39.11

GNU MPのGPGPU対応って、あんまりやってる人が居ないな。

**,,・´∀｀・,,）っ-○○○** · 2012/10/20(土) 09:42:31.34

依存関係があってなかなか並列化しにくいからね。
もちろん並列度の高いコンピュータに適した数式に置き換えればいいんだが
CPU側もAVX, FMAなどでGPGPUの電力効率の優位も揺らぎつつあるからね

**デフォルトの名無しさん** · 2012/11/17(土) 23:26:42.82

C++AMP調べてるとこなんだけど
textureでピクセル間を補間する方法がわからない。
ひょっとして自前でやらないといけないの？
「C++ AMP API は、サンプリングテクスチャのフィルター処理機能を提供されません。」
とはこのことなのかな。

**デフォルトの名無しさん** · 2012/11/18(日) 21:58:01.39

c++ampにfloat_3のmadやfmaが無いね。 short vector typeのmadやsin, cos などの超越関数も用意してもらいたい。 swizzlingはできるみたい。 OpenCLの方がコードをコンパクトに書けるかな。次期版期待してます。

**デフォルトの名無しさん** · 2013/02/17(日) 08:49:52.18

何故かsea islandのISAが公開
ttp://developer.amd.com/tools/heterogeneous-computing/amd-accelerated-parallel-processing-app-sdk/documentation/

**デフォルトの名無しさん** · 2013/02/17(日) 13:26:30.82

いつもに比べると異常に早いがやっぱsouthernislandのリネームなのかね

**デフォルトの名無しさん** · 2013/02/18(月) 13:29:16.68

ディスクリートとしては出なくても
次世代APUのGCNはsea islandなのかね。

ISA見る限り、SIから確かに機能アップされている。
一番大きいのはFLAT メモリアクセス命令の追加か。
cudaのUVAみたいなもの

**デフォルトの名無しさん** · 2013/03/27(水) 13:39:55.44

SL#（えすえるしゃーぷ）とは、GPUで実行されるプログラマブルシェーダーを、超高級言語である
C#で書けてしまうという夢のようなオープンソースのフレームワークである。
http://monobook.org/wiki/SL_Sharp

**デフォルトの名無しさん** · 2013/03/30(土) 11:07:23.56

>>280
まじかよ、ニンジャコンパイラが全部解決してくれるのか？

**デフォルトの名無しさん** · 2013/04/12(金) 18:59:55.71

超高級言語ｗ

**デフォルトの名無しさん** · 2013/04/14(日) 14:30:38.59

抽象化の度合いが高いぐらいの意味合いでいいんじゃないの
実際問題プログラマが意識してGPUに仕事をさせるコードを書かずに勝手に解決してくれちゃう未来像は来るの？

**デフォルトの名無しさん** · 2013/04/14(日) 18:29:31.68

たぶん、デザインパターンと同じくらいの敷居にはなるんじゃねーの？と言うか、デザインパターンとして確立されるんじゃね。

判らない人には判らないけど、自分の中に消化され始めると自然とそう言う書き方になるみたいな。
あとは、フレームワークでそう言う書き方を強制されて、なにも考えずにそう言うもんだと思うパターン。

**デフォルトの名無しさん** · 2013/04/15(月) 02:40:47.34

コア数が100くらいのGPUがあるとして
整数大規模行列の
全要素の和を求めるのってGPGPUで100倍程度の高速化は可能ですよね？

**デフォルトの名無しさん** · 2013/04/15(月) 07:36:47.18

1コアあたりのシングルスレッド能力がCPUと同程度あって、メモリ帯域幅がそれに見合うほど広ければな

**デフォルトの名無しさん** · 2013/04/22(月) 21:00:09.19

>>283
実際、解きたい問題依存だけど、Hadoopのように、プログラミングモデルを固定して制限を付けて縛ることによって簡単な記述で自動的に高速実行ってのは比較的現実的な未来だと思うよ。
Hadoopってのも、書きたい処理をHadoopの"型"に押し込んで書ければ、残りの面倒な問題は全部Hadoopが面倒見てくれますって事だし。

**デフォルトの名無しさん** · 2013/09/21(土) 12:06:20.79

このスレ立った頃のGPUってもうゴミだよな？

**デフォルトの名無しさん** · 2013/09/21(土) 13:04:48.76

OpenCLで書いたプログラムを9600GTとQ9650で動かしたら同じくらいの速度だった。
今のCPUには絶対負ける。

**デフォルトの名無しさん** · 2013/10/02(水) 19:53:46.66

なんか来てた。
ttp://www.phoronix.com/scan.php?page=news_item&px=MTQ3NDU

**デフォルトの名無しさん** · 2013/10/20(日) 14:42:26.77

GPGPUとMantleの関係を語ってくれよ
mantleってなんなんだ？

**デフォルトの名無しさん** · 2013/11/06(水) 09:10:41.13

>>291
DirectXやOpenGLは長い歴史があって
色々なハードで動かす必要があるので
互換性維持のために滅茶苦茶厚いAPIコールを通さないといけないが
MantleはGCNに絞った最適化さえすればいいので
そういうものがないっつーこと
据え置きはAMD一色だから移植も相互に楽になるし、面倒な最適化も最小限で済む
というのが言い分

**デフォルトの名無しさん** · 2013/12/30(月) 01:57:57.57

DirectXやOpenGLやOpenCLだと実行時に
「君（GPU）この機能使える？」
「よし、通れ！」
というプロセスを踏んでるからnVidia、AMD、Intel、PowerVRなど全部のハードウェアでそれなりに動くわけ
これに対してCUDAやMantleは他社対応を完全に無くすことで最初から最適化されたランタイムで動いてる

商用製品でCUDAがOpenCLより人気なのを見れば分かるように、
他のハードウェアで動かない以上のメリットがあるわけ
とくにゲーム機だと全てAMDで、今はAMDがARMと仲が良いからCUDA以上に広まりやすい環境にあるわけ

このスレ的にMantleは用途が違うだろうけど、HLSL/GLSL直書きでGPGPUをやってた層には移行先として十分かなと言った感じ

**デフォルトの名無しさん** · 2013/12/30(月) 03:37:10.21

Mantleって互換性維持のため分厚くなり過ぎたDirectXに対して
低レベルなところを柔軟に叩けるようにするとかそういうものじゃなかったっけ？

Project Sumatraが楽しみだな
Java9ぐらいでparallelStreamにぺぺっと渡すと
OpenCLやってくれちゃうとか最高じゃん
逆に言うとそれぐらい猿でも簡単に扱えないとなっていうか

**デフォルトの名無しさん** · 2014/01/16(木) 15:13:31.68

Mantle.Netはよ

**デフォルトの名無しさん** · 2014/01/16(木) 19:52:34.70

kaveriのFLOPSのDP:SPが気になるんだが、誰か情報持ってないか？

**デフォルトの名無しさん** · 2014/01/16(木) 20:00:22.38

OpenACCってどうよ

**デフォルトの名無しさん** · 2014/01/26(日) 21:12:44.48

最近CUDAを始めたのですが、簡単な計算(大きな配列のベクタ足し算)を
CPUとGPUにやらせると明らかにGPUが遅いです。

具体的には、VC2012+CUDA5.5でコード(http://www1.axfc.net/u/3155195.dat)を走らせると、
計算時間が次のようになりました(Releaseビルド、x64モード)。
CPU→1.51892e-005[s]
GPU→3.7824e-005[s]
一応計算はできているのですが、どうも性能を引き出せていない気がします。
また、コード中でarraySizeを65536にすると実行できなくなるのは何故なのでしょう？
どの辺書き換えればいいのかを教えて下さいお願いします。

ちなみにGPUはGeForce 610M(理論値で141.7GFlops)、
CPUはCore i5-3210M(1コアしか使わない状態なので理論値20GFlops)です。

**デフォルトの名無しさん** · 2014/01/26(日) 21:24:38.81

メモリ転送のオーバーヘッドがあるから、もっと大きな問題じゃないと効果は出ないよ。

**デフォルトの名無しさん** · 2014/01/26(日) 21:54:26.48

>>299
それは知っているのですが、数値を大きくするとすぐにc00000fdでクラッシュするんですよ……。
<<<grid, block>>>もいろいろ弄っているのですが、どうにも効果が得られません。
今試してみたら、arraySizeの値で実行できるのは25600が最大みたいです。

**デフォルトの名無しさん** · 2014/01/26(日) 22:08:07.85

GLSLからOpenCLへの移行を昨日から始めたけど
GLSLより書きやすいのはいいけど最適化を追い込まないととんでもなく遅くなるんだな
GLSLで複雑な汎用計算やらせるのは難解なパズルゲームみたいで嫌になってたけど
結局最適化の手間を考えたらどっちが楽ということはないんだね・・・

>>298みたいな単純な計算ならGLSLだとバグったような速度が簡単に出るから別世界感が凄い

**298** · 2014/01/26(日) 23:14:26.27

>>300からの続きですが、arraySizeをあまり大きくできないので、
ソースを弄って足し算を各100万回行うように改造しました。結果、
Releaseビルド、x64モードで
CPU→16.5872[s]
GPU→5.77132[s]
となりました。ここからFlopsを出してみると、>>298では
CPUが1078.66MFLOPS、GPUが433.164MFLOPSだったのが、
今回はCPUが1975.5MFLOPS、GPUが5677.73MFLOPSとなりました。
理論値からは明らかに小さいですが、少なくともGPUはより活用できているように感じます。

……結局arraySizeを大きくできない問題は解決していません。
ただ、float・int型にしてみると倍(51200)まで設定出来ました。
つまり、流し込むデータは200KBまでは大丈夫ということなのでしょうか？

**デフォルトの名無しさん** · 2014/01/26(日) 23:22:14.60

>>299
GPGPUはメモリ転送のオーバーヘッドがないHSA(Huma)だよな
PCではど重い処理でない限りAMDのHSAがGPGPU処理の主流になるだろうな

**デフォルトの名無しさん** · 2014/01/27(月) 00:23:03.99

>>298
残念なお知らせ。
そのソースコードでは、GPUの演算時間ではなくGPUの呼び出し時間しか計測してないね。
「実際の演算時間」=「内部ブロック数」*(「内部ブロックの呼び出し時間」+「内部ブロックの演算時間」)だとすると、
「実際の演算時間」-「内部ブロックの演算時間」になっているはず。
ブロック数が充分大きければ誤差だけど、内部ブロック数が1のときは激速になってしまう。

まぁ、実際の運用ではCPUとGPUが並列に動作することを期待するからそれでもいいんだけどね。
いずれにしても、CPUぶん回すよりも手っ取り早いと思っていたら大間違いだよ。
それと、CUDAスレも宜しく。

**298** · 2014/01/27(月) 00:47:11.47

>>304
>そのソースコードでは
え！？　……つまり、
普通にtimeGetTimeかQueryPerformanceCounterとかを使えってことなんですか？
それとも、測定する位置が間違っているということなんですか？
>CUDAスレも宜しく
分かりました。次回以降はそちらにレスすることにします。

**デフォルトの名無しさん** · 2014/01/27(月) 08:23:49.16

>>304
何言ってんだ、こいつ？

**デフォルトの名無しさん** · 2014/01/27(月) 21:34:08.90

>>298 arraySizeが大きいと、CPU版すらStackOverflowになるよ。
http://pastebin.com/Av3YzTGs

**307** · 2014/01/27(月) 21:39:43.68

うっかり、166行目を「cudaStatus = cudaSetDevice(1);」にしちゃったので、適当に直しておいて。

**デフォルトの名無しさん** · 2014/01/27(月) 23:30:12.43

ローカルメモリを使う場合って確保しようとした容量が大き過ぎると
グローバルのほうへ確保されてしまうんだよね？
AMDのGCNはどれくらいまでローカルメモリがあるのか分からないんだけど
試行錯誤して調べるしかないのか

**298** · 2014/01/27(月) 23:50:13.09

>>307-308
調査ありがとうございました。そうか、メモリのせいだったのか……
gridsizeの65536制限は知っていたのですが、block・gridでの
分割方法がイマイチよく分かっていなかったので、実コードで
示してくださって助かります。こちらの環境でテストしてみると、
Releaseビルド、x64モードで

> CPU計算時間:0.060652126[s] -> 276.614[MFLOPS]
> size: 16777216
> size_x,y: 262144,64
> blockSize: 256,1
> gridSize: 1024,64
> GPU計算時間:0.034433924[s] -> 487.229[MFLOPS]
> 最大絶対誤差:0.0000000000000000

となりました。>>298より微妙に速くなった程度ですが、
負荷が軽すぎるせいだということは>>302で確認しています。
ちなみにCUDA-Z でこちらのグラボを計測すると、スレッドの次元が1024x1024x64、
グリッドの次元が65535x65535x65535、演算性能は
int32=47.1[Giop/s]・float=94.0[Gflop/s]・double=11.8[Gflop/s]らしいです。

**デフォルトの名無しさん** · 2014/01/28(火) 01:09:12.72

>>307
冗長なOpenCLに比べてやっぱりCUDAはスマートでいいな

**デフォルトの名無しさん** · 2014/01/29(水) 01:59:06.39

OpenCLのclEnqueueNDRangeKernelでカーネルを実行するときに
global_work_sizeとlocal_work_sizeに同じ値(256,256など)を入力すると
何もエラーは返されずにメモリの参照が壊れて？しまいclEnqueueReadBufferで
CPU側で読み取った値が全て0になってしまいます。

これは仕様なのでしょうか？

**デフォルトの名無しさん** · 2014/02/25(火) 21:16:18.98

visual studio 2013でCUDAが使えないからC++AMPでやるお！

**デフォルトの名無しさん** · 2014/02/25(火) 21:43:30.35

>>313
そのためだけにVS2012と2013使い分けてる俺……

**デフォルトの名無しさん** · 2014/04/04(金) 10:44:13.17

openCL始めたお(・∀・)ﾉ

**デフォルトの名無しさん** · 2014/04/15(火) 02:32:13.65

(・∀・)ﾉ CPUの300倍くらいの性能が出たお！
比較したCPUはE2-2000っていうCPU+GPU=APUだけど全くGPUとしての機能をもってないのでガッカリしたお。

**デフォルトの名無しさん** · 2014/04/19(土) 12:16:56.16

(・∀・)ﾉ ALU(IGP)のE2-2000はHD7770の1/50のパワーしかないが並列性はあるようだ。

**デフォルトの名無しさん** · 2014/04/22(火) 04:44:14.02

AIDA64に測定メニューあるよな

**デフォルトの名無しさん** · 2014/08/29(金) 13:33:23.65

AMDとMS，GPU演算用途向けのコンパイラ「C＋＋ AMP v1.2」を発表
http://www.4gamer.net/games/032/G003263/20140828031/

**デフォルトの名無しさん** · 2014/09/12(金) 04:54:39.38

テキスト処理ってGPUで高速化できないものでしょうか
具体的には
Appache Solr
の検索処理が遅いのでなんとか高速化したいのですが

**デフォルトの名無しさん** · 2014/09/12(金) 09:59:47.51

ボトルネックはメモリでしょう。

**デフォルトの名無しさん** · 2014/09/16(火) 05:52:05.01

テキスト処理なんてわざわざGPUでやるよりSSE/AVXでやったほうが億倍マシ

**デフォルトの名無しさん** · 2014/09/20(土) 00:49:25.20

>>319
Linuxで動かすの以外と簡単みたいです
http://d.hatena.ne.jp/niitsuma/20080102/1411114218

**デフォルトの名無しさん** · 2014/09/24(水) 15:13:16.24

OpenCLでプログラム組んでみたけど、CPUとGPUメモリのやり取りがネックになっているのか、思ったよりスピードが出ない

他の人はGPU利用するにあたってメモリのやり取りとか何か工夫している？

**デフォルトの名無しさん** · 2014/09/24(水) 23:08:38.12

そりゃ工夫するだろう。

**デフォルトの名無しさん** · 2014/09/25(木) 13:28:12.29

ごめん、どんな工夫してるか聞いてみたかったんだ

**デフォルトの名無しさん** · 2014/09/25(木) 16:05:41.16

基本はメモリとのやりとりを少なくするって話でしょ
それ以上の個別の工夫を簡単に説明するのは難しいよね
ケーススタディしたいのならそういう本なり文献なり漁るべき

**デフォルトの名無しさん** · 2014/09/25(木) 18:24:41.04

OpenCVのOpenCLバインディングのコードを参考にしたらいいんじゃないのかな

**デフォルトの名無しさん** · 2014/09/25(木) 21:14:08.96

>>324
kaveri使えよ

**デフォルトの名無しさん** · 2014/09/25(木) 22:16:45.04

OpenCLの1.1と1.2に後方互換性ありますか？

**デフォルトの名無しさん** · 2014/09/27(土) 00:53:04.08

>>329
買えたら買ってるよ
メモリの転送の処理が要らなくなったらと思うと幸せな気分になれるよ

**デフォルトの名無しさん** · 2014/10/14(火) 21:40:56.17

kaveriってOpenCL使うとき、コピーせずにポインタ参照で渡していいって解釈でいいの？
最近GPGPUをやりはじめたばかりだから、的外れなことかもしれんが。。

**デフォルトの名無しさん** · 2015/04/12(日) 00:04:42.92

Boost.ComputeあったらC++ AMPいらなくない？

**デフォルトの名無しさん** · 2015/04/12(日) 19:07:28.43

どうだろう？

**デフォルトの名無しさん** · 2015/04/24(金) 08:27:23.88

http://stackoverflow.com/questions/20154179/differences-between-vexcl-thrust-and-boost-compute
一番高速なのはどれ？

**デフォルトの名無しさん** · 2015/05/10(日) 00:43:24.29

vexclを少し使ってみたけど便利だ
あとはC++AMPみたいにradeonのドライバーの
バージョン上がると使えなくなったりしないなら

**デフォルトの名無しさん** · 2015/06/23(火) 13:14:06.30

GPUの行列演算ライブラリってないですか？
具体的には特異値分解できるのを探してます

**デフォルトの名無しさん** · 2015/06/23(火) 13:41:37.08

>>337
機械学習スレで書いてた人かな？
ちゃんと調べてないけどMAGMAなら入ってるかも
http://icl.cs.utk.edu/magma/overview/index.html
{sdcz}gesvd はサポートしてるって書いてある

ただし、GPUカーネル内から直接呼びたい場合は使えないらしい
CPUからカーネル呼び出しする必要がある
（SC14時点の資料）

**デフォルトの名無しさん** · 2015/06/26(金) 11:47:09.05

>>338
ありがとうございます。

**デフォルトの名無しさん** · 2015/07/08(水) 11:17:46.41

最大固有値
最大固有ベクトル
だけを求めたい場合って、

**デフォルトの名無しさん** · 2015/07/08(水) 11:18:15.73

最大固有値
最大固有ベクトル
だけを求めたい場合って、べき乗法が最速でしょうか？

**デフォルトの名無しさん** · 2015/07/10(金) 17:33:26.71

万病に効く薬はないんやで

**デフォルトの名無しさん** · 2015/10/11(日) 13:46:03.27

VS2015のc++amp仕様が変わった？
ＣＰＵで実行するrestrict(cpu)のマイクロソフトのサンプルコードがコンパイルエラーになる。

**デフォルトの名無しさん** · 2016/01/29(金) 12:38:56.49

>>343
C++AMPは終わりやな

**125** · 2016/01/30(土) 08:09:23.91

マイクロソフトの開発ブログで

＞ Is C＋＋AMP dead ？

との質問にレスが無い。

モスさんどこ行った？

**デフォルトの名無しさん** · 2016/06/15(水) 14:41:37.67

test

**わたしはぐぷぐぷ派です** · 2016/12/28(水) 13:24:22.43

は？ごぽごぽに決まってるだろ？って言われた．．．
頭ごなしに言ってくる人って何なんでしょうね

**デフォルトの名無しさん** · 2018/05/23(水) 23:05:05.10

僕の知り合いの知り合いができたパソコン一台でお金持ちになれるやり方
役に立つかもしれません
グーグルで検索するといいかも『ネットで稼ぐ方法　モニアレフヌノ』

MYAIQ

**デフォルトの名無しさん** · 2018/07/04(水) 23:04:08.53

2XM

**デフォルトの名無しさん** · 2018/07/06(金) 12:37:07.14

MYAIQ

**デフォルトの名無しさん** · 2019/09/05(木) 13:16:11.62

CUDAスレはあるのにROCmスレがない