08/08/29 23:27:49
>>546
そういうのはCUDAの最も苦手とするところだ。
全スレッドで同期を取って、代表1スレッドが合計するのが手っ取り早いが遅い。
全スレッドで同期を取って、代表nスレッドがmスレッド分合計してからnスレッド分を合計するのが無難か。
或いは、n個になった時点でCPUに転送してしまう方がいいかもしれない。
549:デフォルトの名無しさん
08/08/29 23:45:31
転送が遅いんだよなあ
550:デフォルトの名無しさん
08/08/29 23:54:56
ストリームを使えば殆ど隠蔽できるよ。
551:デフォルトの名無しさん
08/08/30 00:47:54
>>548
ｎ個分の計算結果を別メモリに退避しておいて1個のスレッドでそれを順次合計するってこと？
552:デフォルトの名無しさん
08/08/30 01:09:07
>>550
kwsk
553:デフォルトの名無しさん
08/08/30 01:18:19
>>551
そのままＣＰＵ側へ転送してＣＰＵで合計しろってことでしょ？
554:デフォルトの名無しさん
08/08/31 05:37:24
すんません、日本語版ドキュメントが公開されたってあちらこちらで書いてあるので
公式サイト探したんですが、なんかないっぽいんですが、これってもしかして削除されたの？
555:デフォルトの名無しさん
08/08/31 05:47:06
日本語のマニュアルが必要なほど内容濃くないぞｗ
556:デフォルトの名無しさん
08/08/31 05:50:28
グローバルメモリを使っても速度に限界を感じたんで、
テクスチャメモリを使ってみたいんですよね
で、そのために日本語ドキュメントも読んでみたかったんすよ
557:デフォルトの名無しさん
08/08/31 07:52:18
テクスチャメモリはグローバルメモリの代わりにはならんと思うが。
共有メモリも使い難いし、定数メモリはデバイスから書けないし。
# 書けないのはテクスチャメモリもそうだけど。
アクセスパターンを見直したほうが医院で内科医?
558:デフォルトの名無しさん
08/08/31 13:22:07
vs2008には対応するんだろうか
559:デフォルトの名無しさん
08/09/01 12:52:26
2.0出たけど対応してなくてがっかりした
560:デフォルトの名無しさん
08/09/01 23:04:36
>>559
俺は結局2005入れた。
561:デフォルトの名無しさん
08/09/03 07:35:45
>>549
もれのやってるのでは転送に3msec、演算に400msecくらいなので
全然オーバーヘッドにはなってないすわ。演算の負荷小さ杉なんでは。
>>556
書き換え不要な定数行列をまとめて__constant__に置いて見たすが
ほんの3%くらいしか変わらなかったすわ。
両方とも、もれのコードがだめな可能性ももちろん有るけど。大有りッスけど。
562:デフォルトの名無しさん
08/09/03 12:19:40
CUDAじゃなきゃ困るって用途がいまいち思いつかない
563:デフォルトの名無しさん
08/09/03 14:50:39
それは、CUDAじゃなく直接GPUを扱うほうがいいということか、AMDのStreamナントカでもいいということか、なんだんだ？
564:デフォルトの名無しさん
08/09/05 04:17:53
>CUDAじゃなきゃ困るって用途がいまいち思いつかない
なら使わなきゃいい
565:デフォルトの名無しさん
08/09/05 09:42:01
100万回以上回るループとか、
何千回単位の二重・三重ループが有るならCUDAｻｲｯｺｫｩﾝｷﾞﾓｯﾁﾞｲｲｲｨｲ
566:デフォルトの名無しさん
08/09/05 12:40:56
>>564
バーカwwwwwwwwwww
頭悪いなお前
567:デフォルトの名無しさん
08/09/05 14:57:45
>>565
確かに
初体験のあの気持ち・・・忘れられない・・・
568:デフォルトの名無しさん
08/09/05 19:25:13
気持ちいのは分かるけど普段使わないようなものだしね
既にあるものを自力でCUDA対応させるのも気力が沸かないしね
倍精度浮動少数が扱えないしねｗｗｗ
569:デフォルトの名無しさん
08/09/05 19:40:59
>>568
CUDA 2.0は？単精度に比べてかなり速度が落ちるらしいけど。
570:デフォルトの名無しさん
08/09/05 19:43:46
ハードの問題だから無理だよ
やってみた
571:デフォルトの名無しさん
08/09/05 19:58:03
>>570
GT200世代じゃないとハードの問題で使えないのか・・・
572:デフォルトの名無しさん
08/09/06 01:17:51
二重ループは兎も角、三重ループとなるとCUDAは苦手だと思うが。
一重は並列にしても、二重ループが残ってしまう。
最近のIntelCPUはループが無茶苦茶高速だから、WoodcrestでOpenMPでも使われたら太刀打ちできなくなる。
573:デフォルトの名無しさん
08/09/06 02:59:57
CUDAをループの自動並列化だと思っていらっしゃるｗ
574:デフォルトの名無しさん
08/09/06 03:22:15
ところで2.0正式版は皆さん安定してる？
うちは、テクスチャmemoryがなんだか変。
エミュで正しく動いているのに実機だと挙動がおかしい。
575:デフォルトの名無しさん
08/09/06 13:32:41
CUDAでクラスが使えないのが痛いな
木構造系のアルゴリズムは並列処理にかなり向いてるし応用範囲も広いのにな
CUDAで無理やりやろうとすると無駄な処理をわざとさせないといけなくなるし
プログラムがむちゃくちゃ汚くて見てられない
576:デフォルトの名無しさん
08/09/06 14:08:13
PyCUDAかCUDA.NETあたりを使ってぜひ感想を聞かせて
577:デフォルトの名無しさん
08/09/09 02:19:48
CudaArrayに、48ビットや24ビットのRGBのデータを入れて、テクスチャにバインドできている方います？
うちではうまくいかないんですよね。
578:デフォルトの名無しさん
08/09/11 13:36:09
CUDAってCPUモードがあるけどドライバが無い環境だと自動的になるの？
579:デフォルトの名無しさん
08/09/11 22:59:45
エミュレーションモードの話かな?
自動的に切り替わるほど融通は利かないよ。
580:デフォルトの名無しさん
08/09/11 23:09:42
dim3 threads(100,1);
method<<<1, threads>>>();
これはいけるんだけど
dim3 threads(100,100);
method<<<1, threads>>>();
ってやると一回も呼ばれないんだけど
何か勘違いしてる
threadIdx.xとyで２次元的に呼び出せるんじゃないの？
581:デフォルトの名無しさん
08/09/11 23:12:59
スレッド総数は512まで(詳細はdeviceQueryを実行するべし)。
従って、100*100は拙い。
582:デフォルトの名無しさん
08/09/11 23:16:39
追記:
その条件だけなら、dim3 threads(ThreadsOfBlock, ThreadsOfBlock)にして
dim3 blocks(100 / ThreadsOfBlock, 100 / ThreadsOfBlock)を追加して
method<<<blocks, therads>>>()するのが定番かな。
勿論、methods<<<100, 100>>>()でもいいけど効率は若干落ちることになりそう。
583:デフォルトの名無しさん
08/09/12 00:23:28
method<<<dim3(100,100),1>>>();
にしたらいけた
もしやブロックって並列処理じゃないの？
584:デフォルトの名無しさん
08/09/12 01:06:40
GPUのプロセッサからあぶれた分は時間軸方向に並列になります。
585:デフォルトの名無しさん
08/09/12 13:31:48
whetstoneとかのベンチマークをCUDAで動かしたいんだけど、
とってきたソースをCUDA環境で動かすには書き換えないとダメなの？
586:デフォルトの名無しさん
08/09/12 13:47:04
sharedメモリなんだけど制限とかあるの？
method<<<dim3(1000,1000),4, 1000*1000*4>>>
とか
587:デフォルトの名無しさん
08/09/12 18:06:38
ガウシアンぼかし３ｘ３を５００ｘ８００のフルカラー画像で約０．３秒だった
8400GSですけど
こんなもん？もう少し早いのを期待したんだが
588:デフォルトの名無しさん
08/09/12 18:45:28
そのアルゴリズムはCPUでやるとどのくらいかかった？
589:デフォルトの名無しさん
08/09/12 18:51:26
ブロックの分け方をいじったら０．１５秒になった
CPUで同じの組むのめんどいです
エミュレーションモードってどうやってやるの？
590:デフォルトの名無しさん
08/09/12 22:29:38
>>589
＞エミュレーションモードってどうやってやるの？
nvccにオプション指定するだけだよ。
あくまでもエミュレーションだから余計に遅くなるけど。
>>587
サンプルで似たようなのなかった?
フィルタ類は色々あったと思うから眺めてみるといいと思う。
>>586
あんたもdeviceQueryを実行する必要がありそうだ。
sharedにそんなに取ると、物理量を大幅に超えるから巧く動いたとしても無茶苦茶遅くなるぞ。
>>585
whetstoneなんて、並列演算に向かないと思うのだけど。
あー、繰り返しを並列にすればいいか。それだったら移植もそれほど難しくはない。
591:デフォルトの名無しさん
08/09/12 23:29:34
デバイス上にRGB(unsigned char)の画像配列を確保してある状態で
これをOpenGLのテクスチャにホストを介さずにバインドして利用出来る？
592:デフォルトの名無しさん
08/09/13 05:38:14
すいません質問です。
CUDAでFFTやってるサンプルコードとかどこかにないでしょうか。
探しても見つからなかったんです。
593:デフォルトの名無しさん
08/09/13 06:58:27
>>592
simpleCUFFT違う？
>>590
「並列演算に向かない処理を並列にすればよくってよ？」
これ最高よね
594:デフォルトの名無しさん
08/09/13 08:25:22
>>592
そのくらい自分で考えれ、そんなに難しい事じゃなじゃん
595:590
08/09/13 08:48:04
>>593
＞これ最高よね
???
>>592
CUFFT使うだけならnvccも要らない。
596:デフォルトの名無しさん
08/09/16 07:06:46
>>590
誰かが「このアルゴリズムは並列化には向かないうんぬん」と言った
アルゴリズム×128本を同時に実行してしまうとか最高よね、という
意味なのではないかな。確かに上司の驚愕を呼ぶね。
597:デフォルトの名無しさん
08/09/16 13:08:08
俺は並列化できそうな新たなアルゴリズムを考案しろという意味かと思った
598:デフォルトの名無しさん
08/09/16 14:18:16
PIの計算だって本来並列には向かない
599:デフォルトの名無しさん
08/09/16 14:20:31
CUDAでデバッグDLLが付属してないようなのだけど
例外処理ってどうやればいいので？
try..catchとか使えるの？
600:デフォルトの名無しさん
08/09/16 15:34:27
無理。
601:デフォルトの名無しさん
08/09/17 01:10:50
多体シミュやりたいなーと思ってたら，本屋で見かけたGPU Gems3に載ってた
でもそれだけの為に買うのもなー・・・３はいつ原版が公開されるんだろー
602:デフォルトの名無しさん
08/09/17 02:51:27
なんかDirectX11で並列演算に対応するのと
11世代のOpenGLの仕様にOpenCLっていうCUDAの類似品が実装されるらしい
たぶんCUDAはこのまま消えていく
603:デフォルトの名無しさん
08/09/17 10:36:55
GPGPUの可能性を示してくれただけで十分だよ
604:デフォルトの名無しさん
08/09/19 00:12:03
カーネルのネストって出来るの？
__global__ void a(){
...
}
__global__ void b(){
a<<<dim3(100,100),1>>>();
}
void main(){
b<<<dim3(100,100),1>>>();
}
みたいな
605:デフォルトの名無しさん
08/09/19 00:50:28
なんか根本的にひどい勘違いしてなくね？
606:デフォルトの名無しさん
08/09/19 00:55:26
>>604
GPU内部からカーネルを発行することは不可能
607:デフォルトの名無しさん
08/09/19 06:46:20
__device__ でプログラミングガイドを検索汁。
608:デフォルトの名無しさん
08/09/19 09:12:17
>>602
CUDAの内部にOpenCLが含まれる構成だからCUDA
609:デフォルトの名無しさん
08/09/19 17:51:43
処理に時間がかかるとGPUが完全停止してタイムアウトでドライバレベルのエラー出すんだけど
一回そのせいで画面全体がぐちゃぐちゃになってWindowsが操作不能にまでなったぞ
非同期実行じゃないんかい
どうすんのこれ？
610:デフォルトの名無しさん
08/09/19 19:18:20
>>609
それは質問なのか？愚痴なのか？
質問だとしたら・・・分かるな？
611:デフォルトの名無しさん
08/09/19 20:43:54
>>609
OSがVistaなら下の資料を参考にレジストリ弄ってみるとか。
WDDM によるタイムアウトの検出と GPU の回復
URLﾘﾝｸ(www.microsoft.com)
612:デフォルトの名無しさん
08/09/20 12:00:52
192コアとか216コアとか240コアとか
使いづらいですよ！漏れの弱い頭がパンクしそうです！
お願いだから隠された力を覚醒させて256コア版出してくだしあ！！！１１
613:デフォルトの名無しさん
08/09/20 19:27:47
いや一回計算方法実装すれば後はそれを使いまわすだけだろ
614:デフォルトの名無しさん
08/09/22 11:58:30
>>612
CUDAをVer2にするんだ。デバイス情報取得APIでコア数が判るようになっているぞ。
# つーか、256コア版の歩留まりが悪くてサブプロセッサ単位で減らして対応しているんじゃないの?w
# PS3のCBEが7SSEなのはそういう事情だそうだし。
615:デフォルトの名無しさん
08/09/26 13:20:34
CUDAってなんて読むの？くーだ？
616:デフォルトの名無しさん
08/09/26 13:21:28
んだ。
617:デフォルトの名無しさん
08/09/26 14:49:14
cubaがキューバなんだから
cudaはクーダだろ・・・常識的に考えて・・・
618:デフォルトの名無しさん
08/09/26 15:18:21
キュ～(＞＜)～だ
619:デフォルトの名無しさん
08/09/26 19:23:46
8400GSだと、h264エンコで実速出ないねぇ
12fpsがやっとだよ
もう少し速いかと思ってたんだけどな
620:デフォルトの名無しさん
08/09/29 01:41:34
ＣＵＤＡを使いたいと思っている初心者です
macのxcodeでもできますか？？
imacで８８００ＧＳです
621:デフォルトの名無しさん
08/09/29 11:36:23
>>619
よりによって、CUDAが動く最底辺の方のGPUを使わなくても……
>>620
NVIDIAの公式サイトが全てなので、そこを読んで判断してください。
つーか、xcodeってなに?
622:デフォルトの名無しさん
08/09/29 20:31:53
>>620
５万払って雪豹もらってください。
動くレベルじゃないって話だけどね。
623:質問です
08/09/29 20:37:44
初心者です。
CUDAサンプルを動かしてみて感じた事なのですが、
HLSL,GLSL,Cg言語それらを使わずに、使った時のような絵が出せるのでしょうか？
CUDAもGPU上で計算しているみたいなので・・・
624:デフォルトの名無しさん
08/09/30 03:09:36
それじゃあ俺も初心者です
625:デフォルトの名無しさん
08/10/01 14:32:24
じゃあおれも
626:デフォルトの名無しさん
08/10/01 23:36:19
>>623
サンプルを見たのなら判ると思うけど、OpenGLはほぼそのまま使えるようですよ。
627:デフォルトの名無しさん
08/10/04 00:48:53
URLﾘﾝｸ(en.wikipedia.org)
これ見るとCUDAにそっくりだな
CUDAをちょっと修正するだけでOpenCLに対応出来そうだけど
628:質問です
08/10/05 00:39:12
>626
OpenGLは使えるのはわかってます。
CUDAでピクセル単位の計算できるのかが知りたいのです（汗
Cg言語を使ってバーテックスシェーダで計算していた処理を
CUDAで実現いてみたら、うまくいったのですが、
フラグメントシェーダでやってた処理が実現できなくて・・・
そもそもできるのかどうか・・・
ってところが知りたいのです。
629:デフォルトの名無しさん
08/10/05 03:26:43
そのためだけにCUDAは作られました
630:デフォルトの名無しさん
08/10/05 03:28:30
つかサンプルにSobelFilterってのがあるだろあれみれ
631:質問です
08/10/06 14:55:42
>>630
SobelFilter見てみました。
見落としてました。
ありがとうございます。
632:デフォルトの名無しさん
08/10/10 21:46:48
日本の公式フォーラムできた
URLﾘﾝｸ(forum.nvidia.co.jp)
633:デフォルトの名無しさん
08/10/10 22:23:43
なんとも直訳のような回答ばかりw
634:デフォルトの名無しさん
08/10/10 22:48:12
直訳なんかしたらまともな日本語になってるわけないじゃないか
635:デフォルトの名無しさん
08/10/10 22:49:17
あの回答がまともな日本語だと思うのか?
636:デフォルトの名無しさん
08/10/10 23:06:59
ああ俺の中で日本語ドキュメントの思い出が醜化されていたようだ
637:デフォルトの名無しさん
08/10/11 00:32:34
悪くない
638:デフォルトの名無しさん
08/10/11 01:03:23
え、俺これ読んですごいwktkしてるんだけど。
サンプル投稿みたいな質問したら、気が向いたらNVIDIAが答えてくれるかも知れないって事でしょ？
639:デフォルトの名無しさん
08/10/11 02:28:44
正直、ここで聞いた方がましだと思う漏れもいる。
640:デフォルトの名無しさん
08/10/11 06:42:53
投稿者： NVIDIA CUDA Team
どんな人 専門家
自信 自信あり
ちょっと面白いｗ
641:デフォルトの名無しさん
08/10/11 07:43:30
少し読んでみたが日本語のあのドキュメントよりはずっと読みやすいよｗ
あれは酷すぎた
642:デフォルトの名無しさん
08/10/12 11:48:23
それでもどう考えても英語のほうが読みやすいけどな。
643:デフォルトの名無しさん
08/10/12 23:45:50
アトミック処理に放射線の危険はありません。:-)
644:デフォルトの名無しさん
08/10/21 20:40:02
質問です。
CUDAを使って、GPU上で計算している部分の一部を、
FBOとCg言語を使って、オフスクリーンで計算させることってできますか？
もちろん、無駄なことはわかっているのですが・・・
645:デフォルトの名無しさん
08/10/21 22:23:14
なんでCUDAのサンプルってコンソールアプリばっかりなんだろう
コンソールアプリじゃないときは
CUT_DEVICE_INIT(argc,argv);
CUT_EXIT(argc, argv);
のargc,argvってとりあえず 0 と nullとか渡しとけばいいの？
646:デフォルトの名無しさん
08/10/21 22:24:53
>>644
日本語でどうぞ。
647:デフォルトの名無しさん
08/10/21 22:27:14
>>645
そもそもCUT_DEVICE_INITはMultiGPU環境でGPUの番号を指定したりするためにあるようなもの。
0とNULLでもいいけど、敢えて使う必要もない。
CUT_EXITに至っては、プログラム終了時にプロンプトを出すためにあるようなもんだ。
# 詳細は、cutil.hを読め。
648:デフォルトの名無しさん
08/10/21 22:33:37
>>647
ありがとうございます
649:デフォルトの名無しさん
08/10/22 00:59:28
>>645
ウィンドウ出してグラフィックだすようなサンプルも多いが。
650:デフォルトの名無しさん
08/10/22 08:56:39
普通にイメージクラスでCUDAを実装してDLL化して使ってるが
DLLだからコンソールだろうがWindowだろうが使えるぞ
でもお前らにはやらない
651:デフォルトの名無しさん
08/10/22 12:41:14
>>694
頭のほうしか見てなかった...
652:デフォルトの名無しさん
08/10/23 14:06:38
ここでロングパス！
653:デフォルトの名無しさん
08/10/26 13:54:35
URLﾘﾝｸ(pepper.is.sci.toho-u.ac.jp)
654:デフォルトの名無しさん
08/10/26 15:14:01
今日からcudaプログラミングを始めようとしている超初心者です
nvidiaのGTX280のピーク性能は933GFlopsだって歌われているのですが、
しかし240のコアで1296MHｚで動作しているのなら、311GFlopsになるはずですよね？
この3倍の数値の差はどこからくるのでしょうか？
655:654
08/10/26 15:16:30
あ、もしかして1サイクルで３つオペランドを消費するような命令があるってことで
しょうかね？
656:デフォルトの名無しさん
08/10/26 17:22:50
maddならmul+addだから2倍なんだけど、3倍なんてあったかな? とこの前から思っている。
657:デフォルトの名無しさん
08/10/26 18:34:50
URLﾘﾝｸ(journal.mycom.co.jp)
madd + mul で3の気がする
658:デフォルトの名無しさん
08/10/26 19:15:55
ベクタ計算じゃないので最内ループの記述は楽々
それでいて300GFLOPS台の性能がでるなんて夢のようですね
659:デフォルトの名無しさん
08/10/26 19:25:03
巧く嵌まればね。書くのは楽だけど、チューニングが大変なのは変わらないわよ。
660:デフォルトの名無しさん
08/10/26 19:50:18
ローコストにCUDAプログラミングやチューニングのコツを掴むのに適したマシンってありますか？
いま持ってるノーパソのGPUがnvidiaならよかったんだが、そうじゃないので
安くCUDAできるマシンを買おうか検討中。「試し」なので自作とか高価なのは避けたい
参考になる話があったら聞かせてくださいまし
661:デフォルトの名無しさん
08/10/26 19:57:15
予算を教えなはれ。
14万でMacBookとか？
後は、Nvidiaチップ内臓の自作で10万切ることも可能
今デスクトップPC不所持で5万の予算だったら無理。
662:デフォルトの名無しさん
08/10/26 20:03:39
意味のある試しプログラミングができればいくら安くてもいい
中古ノートのオンボロで構わないと思っている
奮発しても10万といったところですかね
663:デフォルトの名無しさん
08/10/26 20:05:52
私は一世代前のCore2Duoで8800GT入れているけど9万ほどだったかな。
今だったら、45nmのCore2Duoでメモリ1GB積んで9600GTでも入れれば結構楽しめると思う。
DosPara辺りのゲーム用PCが丁度そんな感じのスペックじゃないかな。
# 都合いいことに、「ゲーム用」はVistaじゃないことが多いしね。
664:662
08/10/26 20:10:12
>>661
10万と書いたが、MacBookの14万というのはちと食指をそそられるな…
当方unixやlinuxはそこそこ扱えるが、Macは十年以上触ったことがないし
開発環境とかどうなっとるんでしょうか
665:662
08/10/26 20:15:35
あ、そうそう
今のノートに買い替えてから使ってないけどモニタとキーボードはあります
モニタったって今時CRTのSXGAですけどね
>>661>>663
ありがとうございます。参考にします
666:663
08/10/26 20:21:45
ちょっとDosPara見てきたけど、XP足しても8万くらいで作れるようね。後は余裕次第かな。
この手のBTOの常で、キーボードとマウスは嫌でもついてくるけど。
それは別として、MacBookで色々苦労したいのならそれはそれでありだと思う。
私なら、ミニタワーで安く済ませた分で、1280x1024を越える大きさの液晶モニタでも買うけどね。
667:662
08/10/26 20:36:02
>>663
やっぱりMacは今でも苦労が多いのかw でもまあひと通り調べてはみよう。
ドスパラのBTOも、後で自分でチェックしてみることにします。
668:デフォルトの名無しさん
08/10/26 20:57:29
bootcampでlinux入れちゃえばOKジャマイカ？
と思ってググッたら酷い、お勧めできない。
669:デフォルトの名無しさん
08/10/26 21:36:37
GTX260は単体で3万円切っているので、工夫すれば10万でデスクトップ組めるかも？？
670:デフォルトの名無しさん
08/10/26 21:40:29
別にWinノートPCを持っているなら、ミニタワーはLinuxでもいいんじゃね？
CUDAをWinで使うとnvccがgccベースだからC++関連の内部関数がめんどくさいよ。
671:662
08/10/26 22:17:28
まあ先に言った通り自作は手控えたいです。トラブルを楽しむほどの余裕が今はないので。
ボード一枚突っ込むくらいですかね。自分の手でやってもいいやというのは。
>>670
VS2008があるので、できればそっちと連携させながらやりたい
（ガワとか作るのはC#が便利なので…）と思っていたんですが、難しいですかね。
cudaが基本gccだというのは判っているので、まあ何ならIPCで繋げばいいやくらいに
テキトーなことを考えていたわけですが。
672:デフォルトの名無しさん
08/10/26 22:43:00
VC向けのプロジェクトテンプレートあったよね
673:662
08/10/26 22:51:53
>>672
ああ、ありましたね
これから風呂入って寝てしまうので、今日はこのへんで。
皆様どうも、いろいろご親切にありがとうございました。
一応、明日以降もときどきこのスレをチェックするようにします。
674:デフォルトの名無しさん
08/10/27 10:21:12
グラボだけ変えればいいじゃん
675:デフォルトの名無しさん
08/10/27 10:23:30
PCIEマシンを持ってないなら
E1200+9400GTあたりで組めば3万以内で組める
676:デフォルトの名無しさん
08/10/27 10:27:01
ちなみにうちのE1200は３Gで動いてたけど
E7200に変えたのでサブにした
３Gあればメインマシンとしても十分だと思うけど
ただマザーはそれなりにOC耐性の高いものが要るので
どうしてもミドルクラスの1万5000円前後になるけどね
677:662
08/10/27 23:23:34
>>675
だいたいそのへんで検討中です。
Windows XPをインストールするとケチっても5万くらいにはなっちゃうようですが
（自宅PCは98SE→Vistaノートへポーンと飛ばしたので、XPを持ってない）ま、それはそれで。
678:デフォルトの名無しさん
08/10/28 21:31:33
強力な浮動小数演算ができる装置を手に入れても、物理屋、ごく一部の化学屋、工学屋の超上位層ぐらいしか使い道が無いぽ。
679:デフォルトの名無しさん
08/10/28 22:09:02
と無知な輩がわめいております
680:デフォルトの名無しさん
08/10/29 03:42:30
実際ない
681:デフォルトの名無しさん
08/10/29 04:44:41
GPU Gemsの3とかに沢山載ってるよ。
暗号化解析とかパケットフィルタリングとか。
682:デフォルトの名無しさん
08/10/29 05:09:26
パケットフィルタの為に２００W近く常時負荷掛けるなんて無理だ
683:デフォルトの名無しさん
08/10/29 08:49:21
＞物理屋、ごく一部の化学屋、工学屋の超上位層ぐらいしか
もともとゲームやCGでさんざん使い倒されているデバイスなのに何を言うのか
CUDAは、ということであれば、たとえばこれを土台にファイナンス系のミドルウェアが作られたら
そっち系の需要が一気に開花するかもしれない
今んとこCUDAのサイトで紹介されている事例はいささか高尚すぐる気がするけどな
684:デフォルトの名無しさん
08/10/29 09:33:00
どっかの銀行でやってるんじゃないか?
685:デフォルトの名無しさん
08/10/29 12:06:23
銀行ってイメージ的に扱うトランザクションは半端なく多そうだけど、それがSIMDではなさそうな。
必要なのはPOWERに乗ってるような十進演算器じゃないの？
686:デフォルトの名無しさん
08/10/29 13:21:47
そう言えば、NVIDIAの営業が得意気に「銀行系には1000台単位でお買い上げいただいてます」とか言ってたな。
687:デフォルトの名無しさん
08/10/29 14:19:44
それ演算用ではなくて表示用だったりして。
688:デフォルトの名無しさん
08/10/30 00:10:25
DoubleFloatのみで対決したら、最新GPU1機 vs. 最新のクアッドコアCPUのどっちが勝つと思う？
CPUはamd64と、SSE等使用時（誤差を考えればむしろこっちと比較するべき？）の両方で予想してほしい。
689:デフォルトの名無しさん
08/10/30 00:12:02
CPU
690:デフォルトの名無しさん
08/10/30 00:23:35
用途による。
考えるのが面倒くさければCPU。
691:デフォルトの名無しさん
08/10/30 02:35:58
メモリアクセスのペナルティがあるから、その観点で比べてもしょうがないよ
692:デフォルトの名無しさん
08/10/30 04:31:20
銀行で使うとしたら、勘定系ではないだろ。
そうではなくて、商品開発やシミュレーションなど、1円2円ずれてもいいような業務
693:デフォルトの名無しさん
08/10/30 15:04:09
銀行とかトラフィックが確かにすさまじいけど
システム改変するコストもすさまじいぞ
数十年に一回やれるかどうかだろ
それに負荷１００％で２４時間なんてカードが耐えられるとは思えない
694:デフォルトの名無しさん
08/10/30 16:15:06
常時負荷100%という状況になっていること自体設計ミス
60～70%が適正
695:,,・´∀｀・,,）っ-●◎○
08/10/31 02:29:13
>>671
マーケティングの人に直接聞いた話では、そのうちそのへんは改善されると思うよ。
あ、機密事項と言われてるので具体的には言えない。
696:デフォルトの名無しさん
08/10/31 17:56:27
最近銀行システムの開発で、6000人のSE集めた超プロジェクト失敗したものねえ。まあ当然だが。
SEが6000人だからねえ。プログラマはもっと多いとかもう想像つかない。
697:デフォルトの名無しさん
08/10/31 21:10:20
TMPGencのCUDA対応版が出たんでインストールしたんだけど
CUDAの項目にチェックできないのは何故・・？
ドライバは178.24でグラボがASUSの8800GTS(640MB)
698:697
08/10/31 21:19:21
スマソ自己解決
g80はダメなんだってねOTL
699:デフォルトの名無しさん
08/10/31 23:34:51
>>698
ｲｷﾛ。
700:デフォルトの名無しさん
08/11/01 00:39:03
G80はストリーム系のAPIが使えないからねぇ。
701:デフォルトの名無しさん
08/11/01 01:49:28
>>696
JRとかの鉄道や、電力といったインフラ系はもっと大きい。
しかし、大きいが故にPJ失敗しまくってる。
人数を増やせば増やす程、集めた人材の質は低下する。
そして頭脳労働の場合、一番質の低い人のレベルに
足並みを揃えなきゃいけなくなるからなぁ。
しかし戦中・戦後に一気に作ったシステムが老朽化して、
銀行どころでなく大規模な改修がどれもこれも必要なのだが。
#mixiで見掛けたよ＞団子の中の人
702:デフォルトの名無しさん
08/11/01 01:54:45
>>662
HP ML115サーバ機に、GF9400GTあたり刺して、
Linux入れたら？ 1CD-Linuxの knoppix for CUDA
なら、最初からCUDA環境が構築済みで、サンプル
も憑いて来るし。
慶應義塾大学泰岡（やすおか）顕治研究室 Yasuoka Laboratory
URLﾘﾝｸ(www.yasuoka.mech.keio.ac.jp)
個人的にはGF8200なM/BのオンボでCUDA走れば、
裸M/BのCUDAクラスタ組もうかと思ってるが、
CPUやメモリの値段を考えると、ML115の方が
安上がりなんだよな。
703:662
08/11/01 06:57:01
>>702
これはビックリ！こんな激安サーバがあるなんて知らなかった…
激安なのにPCI-Expressとかついてて（x16必須な）nvidiaのグラボもちゃんと動く、
ということでゲームの人達にも人気があると…ふむふむ。
ところで素のML115はメモリ512Mなのだけれど
上記研究室のページによるとknoppix for CUDAの推奨動作環境はメモリ2G以上、とある
ML115を使う場合、安いやつを別に買ってきて刺し換えればよろし、ということですね？
（ML115もhp直販だとメモリ増設オプションはECCつきの高いやつしかない…）
ML115が16k、9400GTが9k、2Gメモリも安いのは3k未満、で30kを切りますな。
個人的にサーバ機もAMD64もknoppixも使ったことがないので、
それらの組み合わせとなると微妙に不安だ（笑）が、いずれにせよこの値段は魅力的
大変参考になりました。ありがとうございます。
704:デフォルトの名無しさん
08/11/01 07:57:06
デモ機で借りたTeslaC1060使っているんだけど、ホストCPUがAMDのPhenom。
Xeonに較べて遅い遅い。普段使っているXeonに8800GTの組み合わせの方が早いって何さw
705:,,・´∀｀・,,）っ-●◎○
08/11/01 08:16:53
つまんない質問だけどGTX2xxの人は電源いくらよ。
＋150Wくらいはマージンとったほうがいいと思うよな？よな？
706:デフォルトの名無しさん
08/11/01 10:03:17
なんに対して+150?
GTX280ボード単体での消費電力は236W、GTX260でも180Wクラス消費するからね。
ついでに言えば、補助電源用コネクタもGTX280は6ピン+8ピンの特殊コネクタが必要だし。
707:,,・´∀｀・,,）っ-●◎○
08/11/01 10:25:14
>>706
システム全体で。500W電源以上推奨って言ってるけどじゃあ500Wで安定するかっていうと
信用できねー
708:デフォルトの名無しさん
08/11/01 10:25:55
無理。
709:,,・´∀｀・,,）っ-●◎○
08/11/01 10:28:03
とすると、マシン一式組んで貸し出してもらうのがベストだよな？
よし参考になった。
710:デフォルトの名無しさん
08/11/01 11:59:03
8800GTなら100Wだし、補助電源コネクタも6ピンだけで済むよ。
711:,,・´∀｀・,,）っ-○◎●
08/11/01 12:04:23
うん、俺も8800GTまでなら550Wで余裕といえるラインかなと思っている。
712:デフォルトの名無しさん
08/11/01 14:45:29
GTX280を使うのなら、700Wクラスの電源が欲しいところだね。
713:デフォルトの名無しさん
08/11/02 15:04:39
CUDAは8800以上のクラスで無ければ意味ない。
8500とかはとりあえず走るだけでパフォーマンスは全然駄目。
714:,,・´∀｀・,,）っ-●◎○
08/11/02 17:39:50
8400GSとかになると額面性能でもCore 2シングルコア以下だな。
715:デフォルトの名無しさん
08/11/02 19:59:18
そだ |------､｀⌒ｰ--､
れが |ﾊ{{ }} ）））ヽ､l l ハ
が |､{ ﾊリﾉﾉﾉﾉﾉﾉ）､ l l
い |ヽヽｰ､彡彡ﾉﾉﾉ} に
い |ヾヾヾヾヾヽ彡彡} や
!! /:.:.:.ヾヾヾヾヽ彡彡} l っ
＼＿＿／{ l ii | l|} ﾊ､ヾ} ﾐ彡ﾄ
彡ｼ ,ｪ､､､ヾ{{ヽ} l|l ｨｪ=ﾘ､シ} |l
lミ{ ﾞｲｼﾓ'ﾃ､ミヽ}ｼｨ=ﾗ'ｧ､ }ﾐ}} l
ヾﾐ ￣~'ｨ''': |ﾞ:ｰ.￣ lﾉ/l | |
ヾヾ " : : !､ ｀ lｲﾉ l| |
>lﾞ､ ｰ､,'ｿ /.|}､ l| |
:.lヽ ヽ ｰ_ ‐-‐ｧ' /::ﾉl ﾄ､
:.:.:.:＼ヽ 二" /::// /:.:.l:.:.
:.:.:.:.:.::ヽ:＼ /::://:.:,':.:..:l:.:.
;.;.;.;.;;.:.:.:.＼｀ｰ-- '" //:.:.:;l:.:.:.:l:.:
716:デフォルトの名無しさん
08/11/03 01:57:15
サブノートPCでCUDA動くようにならんかな。
通勤、出張の途中でいぢってみたい。こんな
時でもないと、仕事に直結しないプログラム
組んでる暇無いからなぁ。
>>703
AMDの場合、メモリコントローラがCPUに内蔵なので、
ECCでもnon-ECCでも使える。安い通常のnon-ECCメモリ
1GBx2枚買ってくればOK。ML115はNTT-Xで買えば、
13800円(送料込)。
717:デフォルトの名無しさん
08/11/03 02:19:41
つ N10J
718:662
08/11/03 06:43:24
>>716
回答ありがとうございます。
719:デフォルトの名無しさん
08/11/03 11:13:39
>>716
俺はサブノートでソースは書いてるよ。動作確認は自宅に戻ってからだけどね。
一発で動けば気持ちいいもんだ。
720:デフォルトの名無しさん
08/11/03 11:52:24
質問です
OpenCLが出たらCUDAはお払い箱ですか？
721:デフォルトの名無しさん
08/11/03 12:00:27
>>719
エミュは動いている?
>>720
いいえ、画像処理だけがCUDAの使い道ではありません。
722:,,・´∀｀・,,）っ-●◎○
08/11/03 12:55:07
Apple主導の言語処理系って流行らんだろ。
GPU版Objective-Cだと思え。
ちなみにNVIDIAから補助もらってる俺は仕事につながるって言えるのかな？
723:デフォルトの名無しさん
08/11/03 13:08:35
>>722
その仕事、こっちにくれw
情報少なくて、参ってるんだ。
724:デフォルトの名無しさん
08/11/03 13:10:10
>>722
ObjectiveCは言語仕様からしてクソだったから流行らなかった。
それだけです。
725:,,・´∀｀・,,）っ-●◎○
08/11/03 13:25:26
>>723
メールサポートだけもらってるけどマニュアル落として自分でやったほうが早いしなぁ
俺のほうこそ各ptx命令のレイテンシ・スループットの資料欲しいんだけど。
Intelはそういうのまめに出してくれるから助かるんだが
726:,,・´∀｀・,,）っ-●◎○
08/11/03 13:29:01
YellowBoxだっけ？
WindowsでもMacでも動くアプリケーションが動くフレームワークとか
大風呂敷広げてあれ結局どうなったっけ？
MicrosoftはDX11があるからOpenCLの標準化なんて破談する可能性大
Appleのフレームワークは地雷ばかりで困る。
727:デフォルトの名無しさん
08/11/03 15:35:29
>>725
なんだ、ないのか。NVIDIAの日本法人は、ろくに情報持ってないっぽいんだよね。
728:デフォルトの名無しさん
08/11/03 16:03:03
>>725
ptxは中間言語だろ？
729:デフォルトの名無しさん
08/11/03 17:02:05
>>726
先入観が身を滅ぼすだろう。
730:デフォルトの名無しさん
08/11/03 17:55:12
質問スレッドなので、唐突に質問するわけですが、ごきげんよう
CUDAのSDKに付いてくる
Programming Guide Version2.0の60ページ目の真ん中あたり
For devices of compute capability 1.x,
the warp size is 32 and the number of banks
is 16 (see Section 5.1);
a shared memory request for a warp is split into one request
for the first half of the warp and one request for the second half of the warp.
As a consequence, there can be no bank conflict between a thread belonging to the first
half of a warp and a thread belonging to the second half of the same warp.
が分からない。
何が分からないのかというと、これはShared Memoryの最適なアクセスに関する記述なんだけど、
ワープの中に並列実行できるスレッドが32個あるというのにshared memoryのバンク数は16個しかない。
普通に考えたら２つのスレッドが同時に１つのbankにアクセスするわけで、
思いっきりバンク競合するはずよね？
でも、この記述はバンク競合が起こらないって自信を持って記述されているわけよ
nVidiaの人教えてちょ
731:デフォルトの名無しさん
08/11/03 18:59:44
Half Warp(つまり16スレッド)ずつスケジューリングされるんじゃなかったかな
だからバンク競合は起きない
nVidiaの人では無いが
なら何でWarp=16スレッドとしないんだろう…というのが俺の疑問
732:,,・´∀｀・,,）っ管
08/11/03 20:56:32
中の人いわく
命令レイテンシ隠蔽のためにクロック毎にインタリーブしてるだけだから細かいことは気にすんな
733:730
08/11/03 22:01:55
なるほど～
ワープの正体は16並列と見つけたり
ってことですな！
734:デフォルトの名無しさん
08/11/04 14:28:34
GeForce 9400MってCUDA使える？
735:デフォルトの名無しさん
08/11/04 16:04:29
2.1でサポートできるように頑張ってるけど間に合わないかもしんないって言ってた
736:,,・´∀｀・,,）っ-●◎○
08/11/08 18:21:17
今月中に何かしら動きが・・うわなにをする
くぁwせｄｒｆｔｇｙふじこｌｐ；「’」
737:デフォルトの名無しさん
08/11/08 18:28:53
個人的には1.3世代の1スロット厚のGPUボードが欲しいのだけれど……
# 出ますと言ってた奴はその後連絡寄越さないしなぁ。
738:,,・´∀｀・,,）っ-●◎○
08/11/09 21:55:53
コードの実行時動的生成（分岐除去とかパラメータの定数化とかってレベルで）って
CUDAではいまんとこ無理なんだよな？
Larrabeeが出たらそういう最適化できる部分はXbyak使おうかなと思ってるんだが
っていうか、SPMDじゃないプログラミングモデルまだー？
739:,,・´∀｀・,,）っ-●◎○
08/11/10 01:14:51
>>730-732
に補足。
各SPは最大2issue同時実行なんだけどデコーダは半速。
1SPあたり4スレッドでインターリーブして同じオペレーションを実行するとちょうど命令供給が間に合う構造だな。
1warp＝
16にすると、デコーダは等速か、半速×2にしないといけない。
デコーダの負荷を抑えたかったんじゃないの？
740:デフォルトの名無しさん
08/11/11 21:20:50
CUDAはじめようと思って調べ始めたんだが、
7xxxシリーズはなんで切り捨てられたのか・・・
今週末に9600GT買いに行かなきゃ
741:デフォルトの名無しさん
08/11/11 21:26:36
どうせなら260だか280あたりにしといたら
742:,,・´∀｀・,,）っ-○◎●
08/11/11 22:15:53
電源とかケースとかの敷居高くない？
743:デフォルトの名無しさん
08/11/11 22:33:54
>>740
切り捨てられたんじゃなくて、始めから想定されてない。
URLﾘﾝｸ(journal.mycom.co.jp)
この連載のはじめの方のGPUの進化を追うと、少しは判るかも知れない。
で、どうせならQuadroFX3700をお勧めします。8800GTとほぼ同一仕様でお値段10倍w
744:デフォルトの名無しさん
08/11/12 22:41:03
CUDA-Zなんて便利なものがありました。
forum.nvidia.co.jp
745:デフォルトの名無しさん
08/11/13 21:17:18
それよりレイトレベンチマークのほう、Ｂｉｏ１００％が作ったのか！
SuperDepthとかカニミソとかが蘇ってきたぜ
スレ違い済まん
746:デフォルトの名無しさん
08/11/13 21:20:01
>>745
URLﾘﾝｸ(noridon.seesaa.net)
747:デフォルトの名無しさん
08/11/13 21:58:30
>>745-746
まだ生きてたのか！
PC-98では大変お世話になりました。
そしてブログを読んでみたら、超わかりやすい！
coalescedの意味とか、8/29のエントリみたいなメモリアクセスが遅い理由とかよく分からなかったんだよ。助かった。
748:デフォルトの名無しさん
08/11/15 01:15:52
vista sp1にCUDAをインストールしたいんだけど
ドライバ：○
↓
tool kit：×インストールが終了しない。。
で上手くインストール出来ないんですが、誰かしりません？
強制終了したらアンインストールの項目にtool kitの項目があるのにアンインストールするとerror:5001で失敗しやがるし。。。
最悪
749:名無し募集中。。。
08/11/15 02:28:14
TMpegEncのCUDA対応は4フィルタだけで今のところあまり効果がないみたい
AviUtilのCUDA対応フィルタもパフォーマンス出ないという理由で公開停止
今からでもチャンスありますかね?
750:デフォルトの名無しさん
08/11/15 03:36:03
作りたいなら是非作ってくれ
751:デフォルトの名無しさん
08/11/15 06:15:48
>>748
管理者権限でやっている?
後柱ね。
752:デフォルトの名無しさん
08/11/15 21:32:05
なんかLinux向けのドライバにCUDA2.1入ってるらしーよ
URLﾘﾝｸ(www.nvidia.com)
753:デフォルトの名無しさん
08/11/17 19:18:56
>>738
URLﾘﾝｸ(www.nvidia.com)
スライド87
754:,,・´∀｀・,,）っ-○◎●
08/11/18 17:41:42
>>752
ついに来てしまったか>>736
755:デフォルトの名無しさん
08/11/18 23:45:16
ラジオシティできるソフトってありますか?できればソース付きで...
756:デフォルトの名無しさん
08/11/20 23:06:52
Cg勉強しようと思って調べてたら、CUDAってのもあるんだな。
それぞれできる事って、具体的に何が違うの？
とりあえず7600GTしか持ってないんで、CUDAは使えないんだが、
Cg勉強するぐらいなら、CUDA勉強したほうが圧倒的に良いなら
対応グラボ買おうと思うんだが
757:デフォルトの名無しさん
08/11/20 23:34:27
そういう何処にでも載っていることすら調べられないのならCgにすれば。
758:デフォルトの名無しさん
08/11/21 13:20:39
CUDA 2.1 beta
URLﾘﾝｸ(forums.nvidia.com)
VC++9とDX10インターオペラビリティがやっと
759:デフォルトの名無しさん
08/11/21 21:39:18
cuda sdkのサンプルを実行するとtest failedと出て実行できないんですけど。
環境はos xp 64, quadro FX 4600です。
先ほどnvidiaからドライバとツールとSDKをダウンロードして
インストールしました。ドライバは更新されています。
visual studio 2005も入れました。
760:デフォルトの名無しさん
08/11/22 17:06:06
Teslaを使っているのですが、電源コードを抜く以外の方法で、装置を再起動
する方法はないでしょうか。
バグのあるコードを何度も実行した結果、cudaMalloc()が返ってこない
状態になっています。
761:デフォルトの名無しさん
08/11/22 18:21:05
たわけた質問だと思いますが、お許しください。
NVIDIA製のカードが入っていないPC上で、
nvemulate.exeを利用してCUDAを使用する事は可能なのでしょうか？
実際の処理に使うのではなく、CUDAプログラミングの練習に使うのが主です。
762:デフォルトの名無しさん
08/11/22 18:21:19
>>760
Sシリーズならホストを再起動するだけで復活しませんか？
763:デフォルトの名無しさん
08/11/22 18:48:56
やっぱ大学くらいしかまだ使ってないのかね
764:デフォルトの名無しさん
08/11/23 00:55:06
最近発売された、GeForce9300、9400を積んだＭＢ、
少し前のGeForce8200、8300を積んだＭＢでも実用ではないですが、
CUDAのプログラミングをして走らせる事は可能なのでしょうか？
誰もmGPUでCUDAを使っていないので…
765:デフォルトの名無しさん
08/11/23 16:49:51
みんな何の計算させてるの？
766:デフォルトの名無しさん
08/11/23 16:53:52
株価予測をリアルタイムに
767:デフォルトの名無しさん
08/11/23 19:03:26
株価の予測はできんだろ。アホか。
768:デフォルトの名無しさん
08/11/23 20:24:25
>>767
阿呆丸出し乙
769:デフォルトの名無しさん
08/11/23 20:25:59
株価の予測ができたって言ってるのは、数年前の慶応が出してた論文ぐらいじゃねーの？
770:デフォルトの名無しさん
08/11/23 20:35:29
>>768
予想と予測は明確に違うんだぜ？
771:デフォルトの名無しさん
08/11/23 23:52:44
論点がづれてるー髪もづれてるー
772:sage
08/11/25 12:13:34
>> 762
shutdown -> 電源切断 -> 電源投入の手順を踏むと、復活しました。
ただのrebootで良いかどうかは試していません。
773:アク禁中なので纏めてレス
08/11/25 12:18:30
>>772
色色と掲示板の使い方を間違っているw
で、reboot試してないなら報告しなくていいから。
>>771
髪はずれないと思うぞ、髪は。
>>765
私の所では、最近はFFTWの代わりにCUFFTでFFTを計算させている。
>>764
NVIDIAのサイトのCUDA ZONEでリストアップされていれば、使える。
>>763
んなこたーない。
774:デフォルトの名無しさん
08/11/25 15:29:23
--device-emulationでは正しく動くけれども、実機では動かないときには
ソースコードをにらむしかないのでしょうか。
nvcc --device-debug (-G) というオプションがあったので、これをつけて
コンパイルすると、ptxas が Parsing error を出して失敗します。
URLﾘﾝｸ(forums.nvidia.com)
の会話を見ると、--device-debug は今年5月の段階ではまだ使えなかった
らしく、私の場合と現象が似ているので、以前としてまだ使えないままか
と思ったのです。
775:デフォルトの名無しさん
08/11/25 18:33:51
>>769
そりゃ、予想は出来る罠
ただ、外乱はいつも不明だし、確定解は得られない。
つまり、最尤推定しかできないし、当然推定結果が大ハズレってことも、
初めから推定理論に謳われてる
776:デフォルトの名無しさん
08/11/25 19:24:20
株価予想が正確になればみんなそれを信じて買うようになるでしょ
予測自体が株価に影響を与えだして本来の予測とは違う値動きを始める
そして的中率は下がる
つまり一定以上の正確な予測を行うことは不可能なのだ
777:デフォルトの名無しさん
08/11/25 20:36:28
ここには、当たり前の簡単なことを、必死に難しく言おうとしてる
能無しがたくさん居るようだねｗ
778:デフォルトの名無しさん
08/11/26 02:25:54
１つの.cuの中で実装しているglobal関数の個数によって、
Kernel呼び出しのターンアラウンドタイムが変わるという奇妙な現象に遭遇してます。
特にKernelで処理するデータが少ない時に顕著になります。
関数の数を５～１０個で変えてみると、ターンアラウンドタイムは
最悪値で80μsec、最良で30μsecでした。
この値は
timer.start();
for (int i=0;i<100;i++) test_kernel<<grid,thread>>(test);
cudaThreadSynchronize();
timer.end();
みたいな書き方で調べてます。
9個目、10個目あたりで底があるようなのですが
こういう現象について、合理的な説明はありますか？
僕にはさっぱり見当がつかないのであります。
779:デフォルトの名無しさん
08/11/26 02:28:14
何かを勘違いしている
780:アク禁中（以下略
08/11/26 20:28:00
>>778
再現できるソースを貼ってくれたら解析するじょ。
781:778
08/11/26 23:00:43
#include <windows.h>
#include <stdio.h>
#include <cuda_runtime.h>
//Round a / b to nearest higher integer value
int cuda_iDivUp(int a, int b) {return (a + (b - 1)) / b;}
#define BLOCK_DIM ( 32)
template <unsigned int loops>
__global__ void testcuuuuKernel(float* d_h0, unsigned int size)
{
unsigned int i = blockIdx.x * blockDim.x + threadIdx.x;
if (i < size) {
float d = d_h0[i];
for (int j = 0; j < loops; j++) {d -= j * 0.1; d += 0.9;}
d_h0[i] =d ;
}
}
void dummy() {
dim3 block(BLOCK_DIM, 1, 1); dim3 grid(1, 1, 1);
testcuuuuKernel<4><<<grid, block>>>(NULL, 0);
//testcuuuuKernel<5><<<grid, block>>>(NULL, 0);
//testcuuuuKernel<6><<<grid, block>>>(NULL, 0);
}
int main(int argc, char* argv[]) {
cudaDeviceProp deviceProp;
cudaGetDeviceProperties(&deviceProp, 0);
cudaSetDevice(0);
#define DATA_SIZE ( 100)
unsigned int byte_size = DATA_SIZE * sizeof(float);
782:778
08/11/26 23:02:57
float* data = new float[DATA_SIZE];
for (int i = 0; i < DATA_SIZE;i++) {data[i] = i;}
float* d_data; cudaMalloc((void **)&d_data, byte_size );
cudaMemcpy(d_data, data, byte_size, cudaMemcpyHostToDevice);
LARGE_INTEGER nFreq, nBefore, nAfter; //TIMER初期化
DWORD dwTime;
memset(&nFreq, 0x00, sizeof nFreq);
memset(&nBefore, 0x00, sizeof nBefore);
memset(&nAfter, 0x00, sizeof nAfter);
dwTime = 0;
QueryPerformanceFrequency(&nFreq);
#define LOOPNUM 100
dim3 block(BLOCK_DIM, 1, 1);
dim3 grid(cuda_iDivUp(DATA_SIZE, block.x), 1, 1);
for (int k = 0; k < 10; k++) { //試行の試行
//start!
QueryPerformanceCounter(&nBefore);
for (int i=0; i<LOOPNUM;i++) { testcuuuuKernel<3><<<grid, block>>>(d_data, DATA_SIZE); }
cudaError err=cudaThreadSynchronize();
//stop!!
QueryPerformanceCounter(&nAfter);
cudaMemcpy(data,d_data,byte_size,cudaMemcpyDeviceToHost) ;
dwTime = (DWORD)((nAfter.QuadPart-nBefore.QuadPart) * 1000000 / nFreq.QuadPart / LOOPNUM);
printf("%d usec for %d times kernel launch\n", dwTime, LOOPNUM);
Sleep(400); //ちょっと待つ
}
cudaFree(d_data); delete [] data; getchar(); return 0; }
783:778
08/11/26 23:07:45
再現できるコードを書いてみました。
Dummyという関数でテンプレート展開されている__global__関数の数を調整してみてください。
ちなみに使っているチップはGTX260です
784:アク禁明けw
08/11/26 23:21:32
>>783
面倒だから動かしてもじっくり読んでもいないのだけれど、
カーネル関数はGPUに都度転送することになるから
一回の呼び出し粒度が小さいと転送コストが目立つことになるよ。
その位だと、恐らくは起動コストも無視できないからもっと処理させるべき。
つーか、カーネル呼び出し(<<<>>>)をループで包んだらそりゃ遅いって。
785:778
08/11/26 23:28:05
>>784
もちろんそれは分かるのですが、カーネル呼び出しの処理の内容は、<<<>>>の中の次元数に束縛される
傾向にあると思います。
データ構造などが異なる色々な処理を連携して実行する場合は、カーネル内の分岐では限界があるように思われます。
なので、１回当たりのカーネルのレイテンシを正確に把握しておきたいわけです。
786:デフォルトの名無しさん
08/11/26 23:35:04
＞データ構造などが異なる色々な処理を連携して実行する場合は、カーネル内の分岐では限界があるように思われます。
何にも判ってないと思われ。
787:,,・´∀｀・,,）っ-○◎●
08/11/26 23:39:07
っていうか、分岐ってさ、プレディケートで全部実行するんだよな？
if (cond) { //ここの条件は要素ごとに変わる
funcA();
} else {
funcB();
}
だったら、funcAとfuncBをインライン展開して全部プレディケートつき実行する感じだと思ってるんだが。
788:デフォルトの名無しさん
08/11/26 23:42:08
困ったことに、団子に同意。
789:デフォルトの名無しさん
08/11/27 00:03:15
条件分岐したら負けかなと思ってる by GPU
790:デフォルトの名無しさん
08/11/27 00:32:04
そいえばCUDAって１つのカーネルのサイズが制限されてない？
でかいやつがまったく動かなくて苦労したんだけど
791:デフォルトの名無しさん
08/11/27 01:39:04
どの位かは知らんが、そりゃぁ制限はあるだろうねぇ。
792:,,・´∀｀・,,）っ-○◎●
08/11/27 01:54:09
64Kのコンスタントメモリがあるじゃん。
コンスタントメモリは自分自身では中身の入れ替えは不可能。
あとはわかるよな？
793:デフォルトの名無しさん
08/11/27 02:23:46
cudaで自己書き換えプログラムってできますか？
794:,,・´∀｀・,,）っ-○◎●
08/11/27 02:37:41
GPUのカーネルコード自身で書き換えるって意味なら無理。
PTXのバイナリコードを動的生成とかなら何かやれば可能かもしれない。
その辺の資料を中の人に要求したら
「機密事項ですのでお答えできません」
795:デフォルトの名無しさん
08/11/27 17:49:14
大学の研究室にCUDAプログラミング用のコンピュータが導入された！と喜んでいたら、
HP ML115 + げふぉ8400GSカードだった…orz 学習用仕様で萎えた…
796:デフォルトの名無しさん
08/11/27 17:56:05
まぁ、学部生なら十分だろ。
どうせ大した論文も書かないくせに。
797:デフォルトの名無しさん
08/11/27 18:59:33
>>796
お前みたいなの、必ず湧くよなｗ
人を馬鹿にしたら、自分が偉くなるとでも思ってんの？
798:デフォルトの名無しさん
08/11/27 19:42:03
でも俺もそこまでは言わないにしても
学習用仕様で萎えたってセリフは贅沢だと思うよ。
799:名無し募集中。。。
08/11/27 19:46:05
値段を見て萎えたって意味だろうが性能的には十分な気がするんだけど
800:デフォルトの名無しさん
08/11/27 19:54:22
そもそもGPGPGUはなんなんだ？8800はVGA用途だろ
801:デフォルトの名無しさん
08/11/27 20:23:28
>>797
この程度で反応するなよｗ
社会でやっていけないぞ
802:デフォルトの名無しさん
08/11/27 22:49:14
>>795
それなんて俺
CUDA使ってみたくて買い換えを考えていたけど
PCI Express x16バス搭載で1万だったのでML115と
玄人志向の8400GSカード買ってきてCUDA環境を手に入れたぜ
メインで使ってるマシンより性能がよくてこのままメインになりそうな予感
803:デフォルトの名無しさん
08/11/27 22:51:16
事前に断っておくと俺は批判されるとチンコがたつよ
804:デフォルトの名無しさん
08/11/27 23:04:29
自宅でCUDA使ってる人のコンピュータスペックってどんなもん？
805:デフォルトの名無しさん
08/11/27 23:05:58
>>804
c2d e6320
806:デフォルトの名無しさん
08/11/27 23:11:07
>>804
PenD920 GeForce9600GT
807:デフォルトの名無しさん
08/11/27 23:31:34
>>804
Core i7 920 + GeForce GTX 280
808:デフォルトの名無しさん
08/11/28 00:15:30
>>804
Xeon + GeForce GTX280
809:798
08/11/28 03:29:12
>>799
いや、新技術半可通の俺が察するにGeF8000系は
CUDA対応ハードの底辺だから文句言ってるんだよ。
研究室で導入する(自腹じゃない)んだからもっと良いのよこせってことだろ？
ぶっちゃけ研究といっても今の時期じゃ全体的に未成熟だし
そのスペックでさえ限界動作なんぞせんだろと思って贅沢だなと。
810:デフォルトの名無しさん
08/11/28 03:39:20
>>800
そんなことはない。GPGPUは対応ハードならグラボ1枚でも2枚でも一応使えるんだよ。
実例は押さえてないんでなんともいえないが
競合さえおきなきゃRadeon(VGA)・GeF8800以上(GPGPU専用)ということも出来るはず。
用途と価格の兼ね合いもあるだろうけど、
とりあえず試したいとかコストパフォーマンス的に有利な部分もあるかもしれん。
4gamerの記事が参考になるよ。
URLﾘﾝｸ(www.4gamer.net)
811:デフォルトの名無しさん
08/11/28 04:14:57
>>809
底辺どころかメインストリーム
812:798
08/11/28 06:47:41
>>811
ラインナップ上は確かGeF8世代からの対応だから一応底辺だろう。
>>795の萎える理由に対しての推測だからこの説明でいいんじゃね？
もっとも現状ではSP数の差による性能差がそこまで決定的じゃないから
2x0とかQuadroを期待してるんだったら贅沢なんじゃねって話。
813:デフォルトの名無しさん
08/11/28 07:50:33
GT2x0:性能は兎も角、(電源などの)要求仕様が一般的でない。
QuadroFX:同程度の性能のGeforceの数倍の価格と言う段階で、論外。
# なんて書くと、「アキバ的発想云々」って言われるんだけどねw
もしこれらが理由なら、見識不足と言わざるを得ない。
単に、>795はML115が気に入らないんだろw
まぁ、今時なら8400GSでも8600GTでも9600GTでも大して値段変わらんと思うが。
814:デフォルトの名無しさん
08/11/28 11:07:07
ML115の電源じゃパフォーマンスレンジのGPUは無理ね
815:名無し募集中。。。
08/11/28 11:14:47
ML115の電源はサーバー向けだけあってそれなりに良い電源だよ
電源投入後の全開爆音は凄いけど
ちなみに容量は370W
816:デフォルトの名無しさん
08/11/28 11:52:57
それじゃ精々8600GTSか9600GT止まり。8400GSは案外無難な選択じゃないのかな。
817:デフォルトの名無しさん
08/11/28 13:12:49
ちょっと誰か日本の公式フォーラムのcudaggさんの日本語をなんとかしてあげて
818:デフォルトの名無しさん
08/11/28 13:51:35
Ubuntuの話なら、一応意味は判るでしょ。
つーか、ここでフォローしても意味ないしw
819:デフォルトの名無しさん
08/11/28 18:56:55
>>813
GT2x0で一般的でないとか言ってたらCUDA自体一般的じゃないだろjk
ML115が不服と言うのでも十分贅沢……。俺が行ってた某理系大学なんか(ry
820:流石理系大学出は日本語読むことさえ放棄しているようだな。
08/11/28 18:59:40
>>819
GT280は200Wの電源とそれに見合う通風環境が必要になるってこった。
この際、CUDAが一般的かどうかが問題じゃない。
821:考えること放棄してるやつよりマシだよ。
08/11/28 19:53:01
>>820
CUDAスレでの話だからマシン云々というよりは
単に出たばっかのGT2x0系じゃなかったことに萎えてるのかと思っただけじゃん。
そもそも研究室とかだったら要求仕様が一般的でないなんてことは些細なことだ。
例えばOracle高価だから絶対使いませんなんて真似しないだろ。
そういうところは必要あれば用意するだろうよ。
GT2x0載せるとしたらそりゃML115じゃ無理だろうけど、それに載せろとは誰も言ってないし。
一般的云々言われなきゃ俺だってCUDAが一般的かどうかなぞ持ち出さんわ。
822:デフォルトの名無しさん
08/11/28 19:55:38
グラボの方に萎えてると思ったもうひとつの理由が、
CUDAプログラミング用ということだから
マシン自体のスペックはそれほど高くなくても良いはずなんだ。
電源さえ足りてればな。深読みしすぎた俺が悪かったよ……。
823:デフォルトの名無しさん
08/11/29 01:30:32
そこいらでレスをつけてるやつらの中に、8000系では一部使えない命令がある件を指摘するやつが居ないのは何でだぜ？
お前ら実は妄想だらけでなんもしてないんじゃないの？
824:デフォルトの名無しさん
08/11/29 01:46:30
8400GSならCompute Capability 1.1だから
GT200で追加された命令でもなければ使えるわけだが。
825:デフォルトの名無しさん
08/11/29 02:59:51
>>823
1.0世代は8800GTX,GTS,ULTRAに積まれたG80くらいで、8600GTや8400GSに積まれたG84やG86は
1.1世代だと言う知識もなしにここ見てたの? 他人を妄想だらけなんて指摘してられる状況じゃないじゃんw
826:デフォルトの名無しさん
08/11/29 03:20:49
CUDAを単純に使いたいやつは、
HP ML115 (URLﾘﾝｸ(nttxstore.jp))
と
G98の8400GS買えばおｋだよ。
827:デフォルトの名無しさん
08/11/29 19:48:41
グラボ早い順に並べるとどんな感じ？
やっぱりSP数×クロックなの？
828:デフォルトの名無しさん
08/11/29 21:13:40
仕様を限定せずに一般的に言うなら、そりゃそうだ。他にどんな要素が来ると?
特定の応用でメモリ転送が重視されるのならメモリ帯域やバス仕様にも依存するし、
メモリ量が多い方が高速なアルゴリズムを採用できる応用ならメモリ量にも依存するわけだけど。
あー、1.0世代だとストリームが使えないと言う大きな欠点はあるね。
829:デフォルトの名無しさん
08/11/30 00:11:00
チップ外は考えてなかったφ（．．）ﾌﾑﾌﾑ…
後からSPが減ってクロックが上がった製品が出るじゃない？
だいたい値段は一緒だけど、劇的に性能が違うのかなと…
830:デフォルトの名無しさん
08/11/30 01:46:04
同じ名前のSP減ったのは、恐らく歩留まりの関係で在庫処分しているだけだと思う。
使い方にも拠るけど、クロックが利く応用なら結構変わるかもね。
# SP数が利く応用なら逆に遅くなりかねない。
831:デフォルトの名無しさん
08/12/01 03:44:29
やっぱみんなPDEとか立てたり解けたりできるの？
832:デフォルトの名無しさん
08/12/01 06:49:36
そうだね
SDEとかね
833:デフォルトの名無しさん
08/12/01 23:36:29
>>795
あぅっ、それはウチかも知れない。
まぁどーせ学生の作るプログラムなんて、レポート見てると
自分で作ってるのは15%ぐらいで、残りの人は友人のコピペ
改変か、ググって見付けたページのコピペ改変だからな。
予算の都合ってやつで、1人1台用意するならML115+GF8400
に成ってしまう。許せ。
おっ、これは?!と思うプログラムが出てきたら、GTX280で
走らせてやるから頑がってくれ。
834:デフォルトの名無しさん
08/12/01 23:40:13
>>816
PCI-Ex補助電源コネクタ無いので、GF8600GT
か、GF9500GTまでが無難なところ。
835:デフォルトの名無しさん
08/12/02 00:15:15
texture<float, 1, cudaReadModeElementType> tex;
のテンプレートで
> error C2018: 文字 '0x40' は認識できません。
ってエラーが大量に出るんだが、みんな出ないですか。
オレッスカ
textureのテンプレート使わなければでないのですが。
CUDA 2.1 + VC++ Express 2005
836:デフォルトの名無しさん
08/12/02 00:20:06
>>835
PDFからコピペしてない？
837:デフォルトの名無しさん
08/12/02 00:36:23
>>836
一回行コピペしてエラー出たので消して書いたのですが、
ファイルに変なコードが残るとかありますかね。
でもその行消すとエラーなしで、手書きで書いたら、
↓のようになります。
1>.cu(54) : error C2018: 文字 '0x40' は認識できません。
...
1>.cu(54) : error C2018: 文字 '0x40' は認識できません。
1>.cu(54) : error C2065: 'COMPILER' : 定義されていない識別子です。
1>.cu(54) : error C2146: 構文エラー : ';' が、識別子 'ERROR' の前に必要です。
1>.cu(54) : error C2065: 'ERROR' : 定義されていない識別子です。
1>.cu(54) : error C2143: 構文エラー : ';' が '<template-id>' の前にありません。
1>.cu(54) : error C2146: 構文エラー : ';' が、識別子 'tex' の前に必要です。
1>.cu(54) : error C2275: 'texture<T,dim,__formal>' : この型は演算子として使用できません
1> with
1> [
1> T=float,
1> dim=1,
1> __formal=cudaReadModeElementType
1> ]
1>nv_som_gpu.cu(54) : error C2065: 'tex' : 定義されていない識別子です。
838:837
08/12/02 00:46:22
texture<float1, 1, cudaReadModeElementType> *tex = new texture<float1, 1, cudaReadModeElementType>;
ならOKなので、スタックにtextureを置くとダメな条件があるのでしょうか。
839:837
08/12/02 01:31:47
kernelに引数で渡せないので
スタックにおいたらダメみたいでした。
解決。
840:デフォルトの名無しさん
08/12/02 23:13:52
まだ、初めてみようかと考えている初心者以前のものです。
現状の自作数値解析プログラムはMPIで各ノードの各コアにプロセスを振っております（ノード内もMPIで並列）。
これの自然な拡張としては、MPIの各プロセスがGPGPUを使うという形になるかと思います。
しかし、マルチCPUのノードの場合、ひとつのGPUを共有することになります。
一つのプロセスがGPUのシェーダのうち1/4だけ使って、4プロセスから同時にGPUを使うなんてことは可能なのでしょうか？
841:デフォルトの名無しさん
08/12/02 23:47:13
>>839
テクスチャはコンパイルするとただの数値みたいになるので、
グローバル変数としてそのまま使う以外のことはほとんど不可能、らしいよ。
関数にパラメータとしてを渡すのも、ポインタを得るのも無理。
842:デフォルトの名無しさん
08/12/03 00:02:52
>>840
CUDAを使う場合、4プロセスから同時に使うことも可能。
但し、GPUをどう割り振るかはCUDAドライバの御心次第。
厳密に制御するには、GPU1枚ごとに担当スレッドを設けることになる。
# その場合、プロセス間通信でJOB型にするか1プロセスだけでGPUを占有するかは設計次第。
それはいいけど、プロセス全部分けるとマルチスレッドに較べて効率落ちないかい?
843:840
08/12/03 00:20:11
>>842
ありがとうございます。
やはりリソースの競合が起きるようですね。
マルチスレッドは単純に覚えることが増えるのでやってませんw
スパコンのマニュアルでMPIを覚えたので、WSクラスタでもそのままの手法を持ってきてます。
一応、ノード内では共有メモリを使って通信するようなオプションでmpichをインスコしてるので多少はマシでしょうという感覚です。
844:名無し募集中。。。
08/12/03 02:19:05
NVIDIA、PhysX/CUDAを活用する「パワーパック」の第2弾を提供開始
URLﾘﾝｸ(journal.mycom.co.jp)
845:デフォルトの名無しさん
08/12/03 03:31:23
0x40でググったら
>文字 '0x40' は認識できません。 原因： ソースコード中に全角空白 ' ' が使われています。
まさか、な
846:デフォルトの名無しさん
08/12/04 01:11:01
こんな感じで網羅的に点を回転させるプログラムを書いているのですが、
Thread数が320を越えるあたりでrotate関数に全ての引数が渡らなくなってしまいます。
その時でも計算量が少ないからか、roll_axisのz成分などは、引数として機能しています。
roll_axisのx,y成分も適当な値に変えると引数として働くようになります。
rotate(&coord, &pitch_axis,ANGLE);をコメントアウトすれば,roll_axisのx,yはThread数が多くても引数として渡ります。
計算量が多くなると(特に三角関数？)起きる気がするのですが、何が回避する方法はあるのでしょうか？
--ptxas-options=-vはUsed 42 registers, 56+28 bytes lmem, 2080+32 bytes smem, 8024 bytes cmem[0], 88 bytes cmem[1]て出ます。
-deviceemuでは正常に動作します。
847:846
08/12/04 01:13:18
>>846のソースコードです。
####kernel(Thread１つが回転させる点１つに対応)####
__device__ runDevice～～の一部
float4 yaw_axis = make_float4(0, 0, -1, 0);
for(iyaw = 0; iyaw < limY; iyaw++){
float4 pitch_axis = make_float4(-sin(radian*iyaw), cos(radian*iyaw), 0, 0);
for(ipitch = 0; ipitch < limP; ipitch++){
float4 roll_axis = make_float4(cos(radian * iyaw) * cos(radian * ipitch), sin(radian * iyaw) * cos(radian * ipitch), sin(radian * ipitch), 0);
for(iroll = 0; iroll < limR; iroll++){
if(tid == iroll + ipitch * limR + iyaw * limR *limP)
g_mem[tid] = coord.x; //回転できているか確かめる。
rotate(&coord, &roll_axis, ANGLE); //回転させる点の座標と、回転軸と角度を与える。
}
rotate(&coord, &pitch_axis, ANGLE);
}
rotate(&coord, &yaw_axis, ANGLE);
}
__device__ void rotate(float4 *coord, float4 *axis, int angle){
coord->x = axis->x;
coord->y = axis->y; //とりあえず現段階では引数が渡るか確かめてるだけ
coord->z = axis->z;
}
848:デフォルトの名無しさん
08/12/05 14:25:22
誰か、すげぇ簡単単純だけど、真理を付いてる様な神サンプル晒してよ
849:デフォルトの名無しさん
08/12/05 15:02:45
>>848
どんなの? SDKのサンプルやbioのblog辺りじゃお気に召さない?
具体的なテーマがあって、実装が難しくなさそうなら作ってもいいけどね。
850:デフォルトの名無しさん
08/12/05 21:35:18
ATI Streamから来ました。
スイマセン、場違いな場所に来てしまった。
AもまだなのにCなんて出来ません
851:デフォルトの名無しさん
08/12/05 23:37:57
共産主義者の書いたマンガによると、Aより簡単らしいぞ。
852:デフォルトの名無しさん
08/12/06 00:23:10
最近の若い連中はしょっぱながCだからな
853:デフォルトの名無しさん
08/12/06 19:46:07
cudaってさ
C-daにしておけばギャグっぽくてよかったんじゃね？
854:デフォルトの名無しさん
08/12/07 12:03:03
それならCarracudaの方がｗ
855:デフォルトの名無しさん
08/12/08 22:50:38
teslaの話題が少なくて絶望した！
856:デフォルトの名無しさん
08/12/08 23:03:23
えー、QuadroFX5600のアナログ回路をとっ外しただけの代物の、何を語れと言うのさ。
聞いてくれたら答えるけど。
あ、4桁シリーズは白根。どうせ事情は一緒だと思うけど。
857:デフォルトの名無しさん
08/12/08 23:44:28
>>855-856
このスレの住人は GPGPUの用途にQuadroFX使ってるの・・・？
ていうかそもそもみんな何使っているの？
858:デフォルトの名無しさん
08/12/08 23:58:17
ML115最強伝説
859:デフォルトの名無しさん
08/12/09 00:12:18
Tesla D870 が至高
860:デフォルトの名無しさん
08/12/09 00:59:33
>>857
このスレともう一つのCUDAスレを見ると、QuadroFXのメリットが書かれていると思うが。
# 普通、使うわけないだろって。
8800GTX、QuadroFX5600、Tesla C870の価格を見る限り、Teslaは未だましだと思う。
>>859
D870ってNVIDIA謹製ミニタワーにC870を二枚入れた代物だっけ?
あれだったら同じ筐体のQuadroPlexの方が潰しが利くと思うのだけど。
861:デフォルトの名無しさん
08/12/09 01:51:03
GeForce GTX 280 極上
862:デフォルトの名無しさん
08/12/09 20:16:56
９６００GT買ったらCUDAでエラー出まくりで問い合わせたら
CUDAのようなメーカー付属のドライバで対応してない機能は保障外とのこと
CUDAをやるために買ったのにCUDAが動くことを保障してないなんて
はっきり言って詐欺だよ
863:デフォルトの名無しさん
08/12/09 20:51:59
DELLにVisual Basicで組んだプログラムでエラー出るんだけど？
って聞いてるようなもんだな。
864:デフォルトの名無しさん
08/12/09 20:54:42
>>862
nvidiaのドライバだとどうなんだ？
マジに教えてくれ
865:デフォルトの名無しさん
08/12/09 21:07:30
>>863
CUDA SDKのテストプログラムでエラーが出まくる
実行する度に計算結果が違ってくる
>>862
どういう意味？NVIDIAのドライバ以外にCUDAが動くの？
付属のCDのドライバ以外はNVIDIAの公式ドライバでも保障しないと言われた
866:デフォルトの名無しさん
08/12/09 23:35:15
>>862
一体どこの糞メーカのPCなんだ?
867:デフォルトの名無しさん
08/12/10 23:15:44
>>865
>付属のCDのドライバ以外はNVIDIAの公式ドライバでも保障しないと言われた
それが普通ですよ。販売メーカとしては、自分のところでテストした
ドライバ以外は保証してないよ。
だから公式ドライバ入れる時とかもよく言われるでしょ。
トラブっても自己責任で、ってね。
868:デフォルトの名無しさん
08/12/11 00:00:11
まあ当たり前なのは分かるんだけど、そこんとこ盲点だよ。
付属のバージョンだと根本的にCUDA機能が入ってないからね。
必然的にアップデートしないといけないのに保障外でしょ。
CUDAは素晴らしいとか大々的に宣伝してるけど動くかどうかは運次第という対応だからね。
これが世間に知れたらCUDAなんて誰も相手にしなくなると思うけど。
869:デフォルトの名無しさん
08/12/11 00:16:59
tesla売るためだろ
870:デフォルトの名無しさん
08/12/11 00:19:56
そうだろうね。TESLAは唯一保障された製品だからCUDAやりたい奴はTESLA買えってことなんだろう。
一般人には何の恩恵もない誇大広告だったというわけですか。
871:デフォルトの名無しさん
08/12/11 00:23:48
メーカー純正フロッピーディスク以外使わない奴なんていなかっただろ。
みんなそうやって保証外の物を使ってきているんだ。今更何を言っているんだか。
872:デフォルトの名無しさん
08/12/11 00:27:34
で、そんなふざけた事言うのはどこなんだ？
873:デフォルトの名無しさん
08/12/11 00:27:39
どこかの誰かが同様にCUDAが使えないとなると自分がCUDAを使う意味がない。
だからCUDAは必要なくなるので別に何も問題ない。
誰もが使えないプログラムなんて作っても時間の無駄でしょ。
保障がないってのはそういう事よ。個人的な恨みとかじゃなく。
874:デフォルトの名無しさん
08/12/11 01:57:49
ノートPCなんてNVIDIAが公式ドライバを提供しないんだぜ。
メーカーも提供しないから、アマチュアの作ったものを探してくるしかないっていう。
875:デフォルトの名無しさん
08/12/11 04:03:59
あー、WindowsでCUDA試してないな、うちのMac
876:デフォルトの名無しさん
08/12/11 21:46:41
保証が無いからどうだとかいうのは自作板でやってくれ。
877:デフォルトの名無しさん
08/12/12 02:15:59
>>857
犬板にも書いたけど、JetWay HA05-GTのオンボッボで動きますた。
USBメモリ起動で、薄いCUDA nodeの出来上がりです。
オンボッボももっと沢山SPU載る日はいつかなぁ。
878:デフォルトの名無しさん
08/12/12 18:46:10
【GPGPU】CUDA/ATI STREAM 速度・画質検証スレ
ｽﾚﾘﾝｸ(jisaku板)
879:デフォルトの名無しさん
08/12/12 20:35:23
これからCUDAを使おうと思っているのですが、visualC++でうまくビルドできません。どなたか解決方法を教えてください。
環境はXPx64Edition,QuadroFX570でx64版のバージョン2.0のドライバとToolkitとSDKを入れました。
VisualC++は2005Expressです。
880:デフォルトの名無しさん
08/12/12 20:39:35
すみません、879ですが、ビルドできないのはSDKの中にあるprojectsのsimpleTemplateです。
webで調べてみると皆さん問題なく実行できるようなのですが。。。
881:デフォルトの名無しさん
08/12/12 22:05:56
どんなエラーが出るとかの情報はないの？
882:879
08/12/12 23:26:03
失礼いたしました。下のようなエラーがでます。
nvcc fatal : Visual Studio configuration file '(null)' could not be found for installation at 'C:/Program Files (x86)/Microsoft Visual Studio 8/VC/bin/../..'
>>444さんと同じような問題と思われます。VisualC++のコンパイラとの相性が悪いのでしょうか。詳しい方、よろしくお願いいたします。
883:デフォルトの名無しさん
08/12/13 06:12:02
こんな連中ばっかりだし、クマには生きづらい世の中になったぜ
884:デフォルトの名無しさん
08/12/13 12:41:18
C++もろくに使えない人がいきなりCUDAに手を出すのもどうか
885:デフォルトの名無しさん
08/12/13 14:55:07
なんか64bit環境でもVCのclは32bit版使わないとだめみたいなことCUDAフォーラム（英語のほう）に書いてあったよーな
886:デフォルトの名無しさん
08/12/13 14:55:41
nvccから呼ばれるほうのclのことね＞32bit版
887:デフォルトの名無しさん
08/12/14 01:21:53
>>882
パスが通ってないんじゃね？
888:879
08/12/17 18:34:27
PATHは通っていると思います。消してみると、
nvcc fatal : Cannot find compiler 'cl.exe' in PATH
となりました。platformSDKのclにすると↓のようになりました。
nvcc fatal : nvcc cannot find a supported cl version. Only MSVC 7.1 and MSVC 8.0 are supported
結局、CUDAtoolkitを32bitにしましたところOKとなりました。885さん、886さん、887さんありがとうございました。
889:デフォルトの名無しさん
08/12/17 20:02:14
最適化してregisterの数を減らしたいんですが、
なんか心がけることとかコツってありますか？
890:デフォルトの名無しさん
08/12/17 23:45:04
ptx出力を読め。以上。
891:デフォルトの名無しさん
08/12/18 00:41:24
なんでわからないのでご教授を・・・・
VS 2005にカスタムウィザードを組み込んでパスも通して空の.cuファイルに書き込んだのに
1>LINK : fatal error LNK1181: 入力ファイル '.\Debug\sample.obj' を開けません。
と出るのですが・・・・何か環境設定間違っているのでしょうか？
892:デフォルトの名無しさん
08/12/18 00:49:51
URLﾘﾝｸ(sourceforge.net)
どうぞ
893:デフォルトの名無しさん
08/12/18 01:47:35
>>892
それ使ってるんですが・・・readmeを嫁ということですね
894:デフォルトの名無しさん
08/12/18 03:39:45
ptx出力がどうなるように最適化していけばいいんですか？
895:デフォルトの名無しさん
08/12/18 07:42:38
>>894
>889なら、registerの数を減らしたいんだろ。ptxの出力を読んで少なくなっていればOKだw
896:デフォルトの名無しさん
08/12/20 00:21:41
>>862
NVIDIA、モバイルGPUドライバを配布開始
～ノートでPhysXなどCUDAアプリケーションが利用可能に
12月18日(現地時間)配布開始
米NVIDIAは18日(現地時間)、モバイルGPU用の汎用ドライバを同社ホームページ上で配布開始した。
対応GPUはGeForce 8M/9M、およびQuadro NVS 100M/300Mシリーズ。
従来、NVIDIA製モバイルGPUのドライバは、ノートPC本体のメーカーが提供していたが、
今回よりNVIDIAが汎用ドライバを提供することになった。対応OSはWindows XP/Vista。
バージョンはデスクトップ版の180代に近い179.28 BETAで、CUDAに対応。これにより、PhysXや、
CUDA対応の動画編集ソフト、Photoshop CS4などでアクセラレーションが効くようになる。
ただし、PhysXについてはビデオメモリが256MB以上必要となる。
なお、ソニーVAIO、レノボThinkPad、デルVostro/Latitudeシリーズ、およびHybrid SLI構成の
製品については対象外となっている。
現バージョンはベータ版で、WHQL準拠の正式版は2009年初頭の公開を予定している。
URLﾘﾝｸ(pc.watch.impress.co.jp)
897:デフォルトの名無しさん
08/12/21 00:11:21
なんでvaioは対象外ですか？いじめですか？？
898:デフォルトの名無しさん
08/12/21 07:03:38
vaioとか一部のノートはリコールされてるから
899:デフォルトの名無しさん
08/12/21 19:00:56
それ以前にノートだとメーカーが色々やってるから
一般的にドライバ更新さえノートのメーカー対応待ちだろ。
900:デフォルトの名無しさん
08/12/21 22:30:31
シェーダプロセッサってエラーがあってもゲームでは無視されて
認識出来ないレベルで画面が崩れるだけだから
一見正常に動いてると思っててもＣＵＤＡは動かない人が多いかも
ドライバが対応してるとか以前にハードが用件を満たしてない
901:デフォルトの名無しさん
08/12/24 10:44:05
どこでエラーになっているのかデバッグしろよ
902:デフォルトの名無しさん
08/12/25 15:53:55
ちょいちょいCUDAをいじり始めてるんですが、OpenCLの規格が決まったことで、
CUDAの仕様も大幅に変わってしまうんですかね。
903:デフォルトの名無しさん
08/12/25 18:15:24
MacBookProなんかだと自前でドライバ突っ込めばCUDA使えたんだけど，
VaioやThinkPadだとどうですか？
使っている人がいれば情報求む．
904:デフォルトの名無しさん
08/12/25 18:21:26
5秒制限の条件とか回避の仕方とかがよくわからないのですが
教えてもらえないでしょうか？
905:デフォルトの名無しさん
08/12/26 01:44:56
ThinkPad T61はOpenGL関係以外は使えた。
906:デフォルトの名無しさん
08/12/26 03:48:04
OpenCLはAPIのみだからCUDAとは別にライブラリが用意されるだけだと思うよ
こっちは専用コンパイラも必要ない
907:デフォルトの名無しさん
08/12/26 14:22:18
>>904
Gridを何回も起動するしかないのでは。だから、途中結果をメモリにのこしとかないといけない。
面倒だよね。
908:デフォルトの名無しさん
08/12/26 14:35:42
>>904
Xを使わない、グラフィック機能を使わなければいけるはず。
…何のためのグラフィックカードやねんって感じだが
909:デフォルトの名無しさん
08/12/26 17:45:27
>>907>>908
やっぱりそれぐらいしかないですか。
現状でグラフィック機能止めて使うのは無理なので
Gridの分割を考えてみます。
返答ありがとうございます。
910:デフォルトの名無しさん
08/12/26 18:15:35
Vistaだとレジストリ弄ればもうちょっと長くできるよーみたいなこと言ってたような、言ってなかったような
911:デフォルトの名無しさん
08/12/26 22:25:09
使うグリッド数を制限すればなんとかならなくもないんだが、効率は落ちるんだよね。
つーか、効率重視ならもう一枚挿すしかない希ガス。
912:,,・´∀｀・,,）っ-○◎●
08/12/27 00:47:07
こんなの教えて貰った
URLﾘﾝｸ(compview.titech.ac.jp)
913:デフォルトの名無しさん
08/12/27 05:05:39
どのみちCUDA実行中は画面の描画が止まるから割り込み入れる工夫は必要でしょ
914:デフォルトの名無しさん
08/12/27 08:47:20
描画要求の方がGridよりも優先されているようにしか見えないんだけど。
止まってくれたらどんだけ効率が改善することかw
915:デフォルトの名無しさん
08/12/27 11:52:08
>>912
面白そうだけど、東工大は遠いなぁ。ワープアSEに新幹線は高い。
わがまま言うと、神戸でやって欲しい。
916:デフォルトの名無しさん
08/12/27 14:13:48
大岡山の東工大なら歩いて行けるぜ！
917:,,・´∀｀・,,）っ-○◎●
08/12/27 23:27:30
>>915
六甲台か？
あの急勾配の坂は登りたくないな。
918:デフォルトの名無しさん
08/12/29 05:06:23
>>912
これは紹介レベルだから、たいした内容ではないよ
919:デフォルトの名無しさん
08/12/29 11:44:34
CUDAをクアッドコアで動くように出来ればいいのに
920:デフォルトの名無しさん
08/12/29 12:15:26
>>919
どういう意味?
CUDAをQuadCoreCPUで動かすことなら何の問題もなくできるのだけど。
921:デフォルトの名無しさん
08/12/29 12:44:32
っOpenCL
922:デフォルトの名無しさん
09/01/05 20:09:11
VIPPERの諸君 立ち上がれ！
以前韓国を打ち破ったのは誰か知らない、小学生や、他サイトの管理人がVIPを馬鹿にしている！
全員集結せよ。そして全員で打ち倒すのだ！
もし時間がない人は、この文章をそのまま他のスレに貼ってほしい！
対策本部
ｽﾚﾘﾝｸ(news4vip板)
今こそ Vipperの意地を見せつけるのだ！
923:デフォルトの名無しさん
09/01/08 17:49:55
nc = 100*100
bs = 50
dim3 dimBlock(bs,bs)
dim3 dimGrid(sqrt(nc)/dimBlock.x,sqrt(nc)/dimBlock.y)
kernel<<<dimGrid, dimBlock>>>(idata, odata, sqrt(nc)
__global__ void kernel(float* idata, float*, odata, int nc)
{
index=blockIdx.x * blockDim.x + threadIdx.x +
+(blockIdx.y * blockDim.y + threadIdx.y) * nc
}
この時のイメージは、Gird:2x2、Block:50x50でよいのでしょうか?
それとこのままグローバルメモリで計算するのはできるのですが、一旦
シェアードメモリに退避して計算してグローバルメモリに戻す方法が
サンプルを見てもうまくいきません。どういう感じになるのでしょうか?
924:デフォルトの名無しさん
09/01/08 18:11:37
いつも疑問に思いつつ使っているのですがグローバルメモリ使うときも
共有メモリ使うときも
float* data;
CUDA_SAFE_CALL(cudaMalloc((void**)&data, sizeof(float) * 100)
CUDA_SAFE_CALL(cudaMemcpy(data1, h_data, sizeof(float) * 100, cudaMemcpyHostToDevice))
になるのでしょうか?これはグローバルメモリにとっているのですよね。
共有メモリの話は、kernelのほうで使用するかどうかですよね?
コンスタントメモリ使うときだけどうも
CUDA_SAFE_CALL(cudaMemcpyToSymbol(data, h_data, sizeof(float) * 100))
としてホストからコピーするんですかね?
925:デフォルトの名無しさん
09/01/08 23:27:14
>>924
cudaMemcpyToSymbolは定数メモリの他、グローバルメモリでも使えるよ。
要は、device空間で定義されたメモリと言うニュアンスでしかない。
逆に言えば、cudaMemcpyToSymbolを使わないとhost側のメモリ空間に割り付けられたアドレスを利用しようとしてしまう。
# この辺が判り難いのよね。
共有メモリはグリッド実行中しか使えないからhost側からは為す術がない。
>>923
どのサンプル見てどうやったらどう巧くいかないのか具体的に。
少なくとも、>923のような処理なら共有メモリの使いようがないんだけど。
926:デフォルトの名無しさん
09/01/09 01:50:13
warp単位32スレッドで同じ命令を行って、
実際に一度に動くのは8スレッド
こういう考え方でいいのですか？
927:デフォルトの名無しさん
09/01/09 18:21:54
>>925
結局現状では、cudaMemcpyXXXの違いは理解できそうにないです。
まあそれはよいとして(全然良くないですが)
>>923
のidataは、実は、一次元配列なんです。なのにカーネルには2次元で利用しよう
としているんです。
float host_idata[nc]
float* idata
cudaMalloc((void**) &idata, sizeof(float) * nc)
cudaMemcpy(idata, host_idata, sizeof(float) * nc, cudaMemcpyHostToDevice)
こんな感じのグローバル変数がidata。
としたとき、ncとbsの設定値によってはうまく処理できないことがあるのです。
これが謎です。わかる方いませんか?もちろんncはsqrtで整数になる値、設定する
値は、割り切れる値です。nc=1000*1000,bs=500で処理結果がおかしくなっています。
kernelの読んだあとエラーでinvalid configuration argument.になってます。
それと、共有メモリのほうですがこれは分かりました。
kernel<<<dimGrid, dimBlock, sizeof(float) * nc>>>(idata, odata, sqrt(nc))
第3引数の共有メモリのサイズを定義しないとだめだった。ちゃんと説明書を
読んでいなかったというレベルでした。これは解決です。
928:デフォルトの名無しさん
09/01/09 19:05:25
>>927
ブロックは512x512x64と出ているけどいちお3次元は分からないので512x512
としてxで512使い切るとどうもyではもう1しかとれないようです。
ncは、そのままだとエラーが出ないレベルは、bs=20つまり20x20=400ブロック
使用。21,22は割り切れないので設定できない。みたいです。
とすると一次元配列を使用するのにわざわざ、x,yの次元のブロックを使いきる
テクニックは必要なさそうですね。と推察できると思います。結果からですが
ただ、これがバグであるなら次のCUDAのバージョンアップで使えるようになる
かもですね。
929:デフォルトの名無しさん
09/01/09 20:53:02
>>928
次元とは別に1ブロックあたり512スレッドが上限じゃなかったっけ？
どっかにきちんと書いてあったはずだからバグではないよ。
930:デフォルトの名無しさん
09/01/09 22:23:41
私は1次元でやってしまうことが多いなぁ。
で、スレッド数についてはdeviceQuery参照で。
931:デフォルトの名無しさん
09/01/10 21:32:30
スレッドの最大数はAPIで取得できる。全部512だと思う。
ブロックの最大は65535だから65535*512以上は1次元のグリッドでは扱えない。
あと最大は512だけど、192か256くらいのほうがパフォーマンスが出る。
932:デフォルトの名無しさん
09/01/14 01:57:36
CUDA 2.1 Release
URLﾘﾝｸ(forums.nvidia.com)
933:デフォルトの名無しさん
09/01/14 21:34:37
情報サンクス。早速、Linux版ゲットした。
934:デフォルトの名無しさん
09/01/15 01:08:35
9600GTでVideoDecodeのサンプル動かそうとしたら正常に動かなくて、
結局SDK2.0+178.24まで戻さないとダメだった。
2.1にもソースは付いてるけど、2.0のままみたいだね。
935:,,・´∀｀・,,）っ-●◎○
09/01/15 07:55:42
>>931
1 warp = 32 threadで、GeForce 8/9が24warp/block、GT200で32warp/blockが最大だから
768thread/blockと1024thread/blockが最大なんだけどね。本来は。
CUDAのドライバ側が512でリミッタかけてるんだ。罠としか言いようが無い。
逆に言うとCUDAを経由しなきゃ目いっぱい使えるかもね。
ただ、スレッドインターリーブすると１スレッドあたりで使えるレジスタ本数が減っちゃうんだよね。
メモリレイテンシを隠蔽するならスレッドを目いっぱい使ったほうがいいし
逆に一時変数を何度も再利用する場合は、thread/blockを減らして１スレッドあたりの仕事量を増やしたほうがいい。
936:デフォルトの名無しさん
09/01/15 15:42:19
CUDAってシーユーディーエーって読んでいいの？
それともクダとか、なんか読み方あったりする？
937:デフォルトの名無しさん
09/01/15 15:57:04
>>936
クーダ
938:,,・´∀｀・,,）っ-●◎○
09/01/15 15:58:34
ネイティブの人は「クーダ」って発音してたよ。
Cellは「くた」だから注意な。
939:デフォルトの名無しさん
09/01/15 16:01:48
ありがとう！明日から会社でクゥゥゥウーダ！！って叫びまくるわ
940:デフォルトの名無しさん
09/01/15 16:02:44
きゅーだ って発音してる俺は異端ですかね？
941:デフォルトの名無しさん
09/01/15 16:23:46
>940
そんな人いたんだ。
942:デフォルトの名無しさん
09/01/15 16:29:33
英語的に クー よりも キュー だと思わね？
943:デフォルトの名無しさん
09/01/15 16:56:19
>>941
異端だな
944:デフォルトの名無しさん
09/01/15 16:57:46
どっちでもいいんじゃね
ちなみにnudeはニュードとヌードの両方の発音が有
945:デフォルトの名無しさん
09/01/15 17:13:01
キュもクも似てるし、まあ仮に間違って言ったとしても気づかれないんじゃない？
946:デフォルトの名無しさん
09/01/15 19:38:28
ホムペにはクーダと嫁
と書かれてたはず
947:デフォルトの名無しさん
09/01/15 22:24:41
"The cuda and my wife"
948:デフォルトの名無しさん
09/01/15 23:40:36
たしか英語版のfaqには
kuh-da
と書いてあったはず
949:デフォルトの名無しさん
09/01/15 23:53:57
2.1はJITサポートがミソ？
950:デフォルトの名無しさん
09/01/16 14:10:31
2.1をつついた人に質問
2.1ではカーネルに渡す引数でポインタ配列は使えますか?
951:デフォルトの名無しさん
09/01/16 18:48:04
共有メモリからグローバルメモリに連続領域をコピーするにはどうすればいいでしょうか
カーネルから呼び出せるmemcpyみたいなのがあれば教えてください
952:デフォルトの名無しさん
09/01/16 18:58:33
cudaMemcpyが長いようで、the launch timed out and was terminated.が
でてしまいす。これなんとかするほうほうありましたっけ?
953:デフォルトの名無しさん
09/01/16 21:53:12
>>952
転送量を減らす。
>>951
cudaMemcpy
954:デフォルトの名無しさん
09/01/17 00:16:47
>>951
スレッド内で連続的にメモリにアクセスするようなコードは望ましくない。
coalscedにアクセスするとなると、自分でコードを書くしかない。
共有メモリから転送を始める前に、同期を取ることも忘れずに。
>>952
参考までに、どんなハードウェアの組み合わせでどれくらい転送しようとしている?
8800GT辺りだと転送できるだけ全部転送してもタイムアウトしないのだけれど。
955:デフォルトの名無しさん
09/01/17 02:52:44
>>954
カーネル実行後の結果データコピーでエラーになっているのですが
sizeof(float)*1000バイトを100回連続でコピーしています。
100個とも別のアドレスに結果をコピーしたいためそうしてます。やはり
いっぺんにメインメモリにいったんコピーしてからやったほうがいいので
しょうか?
956:デフォルトの名無しさん
09/01/17 03:03:49
>>955
sizeof(float)*10000バイトを100回連続コピーでした。
957:デフォルトの名無しさん
09/01/17 07:16:40
>>954
＞ >>952
＞ どんなハードウェアの組み合わせ
958:デフォルトの名無しさん
09/01/17 07:19:52
>>956
同一データ部分を100箇所にコピー？
独立の100箇所をコピー？
959:デフォルトの名無しさん
09/01/17 07:55:29
>>956
それ、本当にメモリのコピーでタイムアウトしてる？
カーネル実行後に同期取らずにすぐにメモリのコピー始めて
100回コピーのどこかでカーネル実行がタイムアウトしてるとか。
cudaMemcpyの戻り値は
Note that this function may also return error codes from previous, asynchronous launches.
ということだし。
960:954
09/01/17 08:44:25
>>955=956
えーと、都合数MB程度ってことか。それだったら>959の言うようにカーネル関数のタイムアウトじゃないかな。
DEBUGつきでCUT_SAFE_CALL(間違ってたら失敬)を使うか、自前でcudaThreadsSynchronize()を呼んで戻り値を見てみて。
カーネル関数起動後は、復帰を待たずにホストの処理に戻るからカーネル関数自体のエラーはカーネル関数起動では捕捉できないのよ。
で、まさかとは思うけど同一データを100回コピーならバス越しにやってはダメ。別のデータならホスト側の用件次第かな。
# でも多分、一旦ホスト側に転送してから分配した方が速い気がする。CPUとメモリ周りがよっぽど遅くなければ。
>>957
下手な突っ込みはお郷が知れるよ。この場合、CPUが遅いんじゃないかとかバスがx1なんじゃないかとかGPUがモニタ表示か何かで
よっぽど負荷が掛かっているんじゃないかとかGPU側メモリが巨大なSTELAなのかとか、組み合わせ次第で色々条件が変わってくるのよ。
961:デフォルトの名無しさん
09/01/17 09:52:40
>>960
＞ 下手な突っ込みはお郷が知れるよ。
質問者が答えてないのを指摘しただけなんだが…
誤解させてすまんかった
962:デフォルトの名無しさん
09/01/17 11:09:08
そういう意味か。あの書き方じゃぁ、誤解されるよ。
まぁ、大量転送じゃないと言う意味で、今回は関係なさそうだね。
963:デフォルトの名無しさん
09/01/17 15:38:06
>>952の話
>>957
ATIの統合チップセットでオンボードのATIビデオカードを無効
PCI-E 1.0 x16バスにGeforce9600GT 512MB
CPUはAMDデュアルプロセッサ(結構遅いやつ)
964:デフォルトの名無しさん
09/01/17 15:55:29
>>952の話
>>959
カーネル実行
a:cudaのエラーチェック(カーネルのエラーのありなし)
cudaThreadsSynchronize
for(100回
結果データをデバイスからホストにコピー
b:cudaのエラーチェック(直前のcudaMemcpyのエラーのありなし)
)
bのとこでthe launch timed out and was terminatedが出てるんです。
先週は時間がきてあきらめて帰ったのでここまでです。
言われたようにcudaThreadsSynchronizeの戻り値見てみたほうがよさそう
ですね。
965:デフォルトの名無しさん
09/01/17 18:45:09
>>951
共有メモリはカーネル起動時に動的に確保されるメモリ領域だから
カーネルが終了したら消えるし、１つの実行スレッドから全部コピーなんて論理的におかしいし
共有メモリはあくまで高速動作と同期が取れる作業用のメモリとして考えたほうがいい
共有メモリを使って計算した結果は１個づつ普通のグローバルメモリに入れてやるのが正しいやり方
>>952
10000を一度に転送して実行しても
1を10000回繰り返して転送しても
実行時間は大差ないんですよ
CUDAで実行する部分は出来るだけコンパクトにまとめて
呼び出す本体の方で特殊な演算関数くらいの感覚でループさせて使うのが正解
966:デフォルトの名無しさん
09/01/18 23:59:40
JCublasについて
おしえてエロイ人
967:デフォルトの名無しさん
09/01/19 11:04:50
>>925の話
cudaThreadsSynchronizeの戻り値をチェックしたら
the launch timed out and was terminatedが出ていました。結果コピーで
エラーで落ちてたのだけどエラーデータは、その前に実行していた
cudaThreadsSynchronizeの問題だったようです。
(cudaThreadsSynchronizeが正常になるまで待つとしても配列を
100x1000回、100x10000回、回すと1分待っても同期とれないようです。)
つまりカーネルに同期できない処理が書かれていたのが原因だと思います。
NVIDIA_CUDA_Programming_Guide_2.0.pdfのp34-35のようなコードで
かつそこのカーネルのdevPtrを参照するだけでなく書き戻す処理をや
っているのでそれがだめなのだと思います。恐らくこういうピッチ処理
の場合は、参照のみが許されるのでは?と思います。
問題のコードをこのpdfの変数に直して下記に書いておきます。
__global__ void myKernel(float* devPtr, int pitch)
{
for (int r = 0; r < height - 1; r++) {
float* row1 = (float*)((char*)devPtr + r * pitch);
float* row2 = (float*)((char*)devPtr + (r + 1) * pitch);
for (int c = 1; c < width - 1; c++) {
row2[c] = row1[c]とrow1[c + 1]とrow1[c -1]やらの計算
}
}
}
このp34-35を見て思うのはindex使わずにこんな処理書いてGPUでやる意味
あるの?と思いますが。自分で書いてなんなのですが
968:デフォルトの名無しさん
09/01/19 17:28:20
お前はあほかｗ