22/10/08 13:36:18.12 d409kTqm.net
.cuファイルで何とか普通のメモリと
GPUのVRAMを同居させて使えるなら
容量の大き過ぎる計算でもスピードは落ちるが
普通のCPUだけの時より速く計算できる、
とか出来たらいいんだけど、まず無理なんだろうなあ
(沖電気はそういうのを上手くやったらしいが詳細が分からない)
301:デフォルトの名無しさん
22/10/08 13:53:17.87 XEAL3BhY.net
何を問題視してるのかよくわからんけど
例えばmkl入れるとかじゃ駄目なのか？
302:デフォルトの名無しさん
22/10/08 14:37:57.21 TKlSmRLn.net
>>300
LinuxならUnifiedMemoryでできるな。
303:デフォルトの名無しさん
22/10/08 16:09:18.26 j9nJcF5K.net
>>302
それは凄い！！！
304:デフォルトの名無しさん
22/10/09 00:49:49.72 KNQys/Sq.net
ホスト側のメモリをピンして使うことはできるけど多分そういう用途ではないよなあ
305:デフォルトの名無しさん
22/10/09 08:33:30.58 DpFtFESu.net
>>300が言ってるのはOversubscriptionだろ。
306:デフォルトの名無しさん
22/10/26 02:06:11.16 XY9sqarF.net
C#でCUDAを使おうとして悪戦苦闘してようやくこのスレにたどり着いたのだが・・・
GPUで計算した結果をCPUに取り出すには結局どうやったらいいんだ？
検索してもサンプルプログラムはほとんど出てこないし、GPU動かして終わりでその結果を使うやり方が示されてない。
教えろくださいお願いしますだ
307:デフォルトの名無しさん
22/10/26 05:14:53.09 2ajidUUz.net
CUDAでdll作成してC#に読み込ませる
こんだけ
308:デフォルトの名無しさん
22/10/26 10:08:16.86 Gl6HUSuY.net
>>274
C言語で普通にCuda使うときだったら、GPU側にメモリ確保してGPUで計算した結果をそこに書き込む。
GPUでの計算が終わったらGPU側のメモリをCPU側のメモリにコピーするみたいな感じだ。
後unified memoryとかいうのでGPUから直接CPU側のメモリに書き込めるらしい。
C言語だったらCUDA sdkにサンプルコードがたくさん入ってるだけどね。
どういう関数使えばいいかはCUDAのドキュメントをちゃんと読めば見つけられるでしょう。
309:デフォルトの名無しさん
22/10/26 10:09:44.08 Gl6HUSuY.net
↑のは>>306への返信です。
310:デフォルトの名無しさん
22/10/26 10:49:12.85 XY9sqarF.net
ありがとうございます。
昨日はとりあえず↓を参考にして以下のように書き換えてみたんですが、これじゃダメってことですよね？
（textBox1の結果は0のままです）
double temp = 0;
var gpu = Gpu.Default; // GPUインスタンス取得
gpu.For(0, 10, i =>
{
temp = 10;
});
gpu.Synchronize(); // ここで同期する
textBox1.Text = temp.ToString();
Alea GPUライブラリを使ってC#で簡単GPU並列プログラミング ? Crow's eye
URLﾘﾝｸ(kzmmtmt.pgw.jp)
Alea GPUで簡単C# GPUプログラミング - Qiita
URLﾘﾝｸ(qiita.com)
311:デフォルトの名無しさん
22/10/26 11:05:33.68 2ajidUUz.net
Alea GPUなんて知らなかったな
で、もう使われてなくね？
それでやる必要ある？
312:デフォルトの名無しさん
22/10/26 11:15:01.82 XY9sqarF.net
GPU使って並列計算できるってのを昨日知った初心者なんで、すいません。
「gpu 並列計算 C#」で検索してもこれしか出てこないんです。
313:デフォルトの名無しさん
22/10/26 11:21:28.43 2ajidUUz.net
で、あればGPUの使い方、CUDAの使い方を、まず勉強した方が良いのでは？
まだGPUとホストのメモリ間のデータ転送とか、さっぱりわからないよね
314:デフォルトの名無しさん
22/10/26 11:45:49.61 Gwv5fEeF.net
cudaは別にコンパイルしてC#から呼べば？このほうが調べやすそう
315:デフォルトの名無しさん
22/10/26 12:52:17.28 XY9sqarF.net
趣味でやっているだけなので調べてすぐにできればいいと思ったんですよ。
>>310でGPU側で計算した結果を渡すだけなのにそれを乗せてるウェブサイトが見つからないなんて、そんな手間な作業なの？
316:デフォルトの名無しさん
22/10/26 13:29:27.83 Gwv5fEeF.net
>>315
cpuとgpuで別々のメモリを見てるんで >>308 のようなことが必要です
317:デフォルトの名無しさん
22/10/26 16:16:55.85 Gl6HUSuY.net
CUDA本体はC言語かC++言語で使う前提なのでそれらの言語ならサンプルコードや情報がたくさん見つかるんだけど。
C#からCUDA使うとなるとマイナーなライブラリを使うことになって情報を見つけづらいかもしれない。英語の情報までくまなく調べればそうじゃないのかもしれないが。
318:デフォルトの名無しさん
22/10/26 16:57:54.98 2ajidUUz.net
C#から使う場合は上にも書いた通り、cuda でdllを作成して
C#からはDllImportで読み込む
319:デフォルトの名無しさん
22/10/26 17:35:34.36 XY9sqarF.net
そうなんですね。ありがとうございます。
ライブラリが用意されているのでそれでできないのなら何のためのライブラリなのかと素人的には感じてしまいました。
320:デフォルトの名無しさん
22/10/26 21:01:13.64 8mk+cARY.net
出来ないかどうかはちゃんと調べないとわからないだろう
簡単に出来るかどうかはライブラリは保証しないよ、特にC#とCUDAみたいな組み合わせだったら。
めちゃくちゃ沢山のグルーコードをC++/CLIで書かなきゃいけないのを省略してくれる、ぐらいなもんでCUDAに対する理解なしに使える代物じゃないと思うけどな。
321:デフォルトの名無しさん
22/10/26 23:22:43.17 XY9sqarF.net
すいません、↓にテストコードが載っていたのでそれを動かしたところ、GPUの計算結果を取り出せました。
お騒がせしました。
Alea GPUで簡単C# GPUプログラミング - Qiita
URLﾘﾝｸ(qiita.com)
322:デフォルトの名無しさん
22/11/26 11:57:28.57 BVUhPWJi.net
CUDA初心者です。RTX2060 superでCUDA環境構築したいのですが、どうしてもcuda.is_availableの結果がFalseとなってしまいます。(Nvidiaのドライバ、Python・CUDA・Pytorchのバージョンを新しくしたり古くして見たり、CUDNN入れてもダメでした。)
python -m torch.utils.collect_envで読み込みした現在の環境は下記の通りとなります。Trueとさせる方法がわからず、もう自分では完全に手詰まりとなっておりますので、ご教授頂けますと大変有難いです。
C:\Users\●●●>python -m torch.utils.collect_env
Collecting environment information...
PyTorch version: 1.13.0+cu116
Is debug build: False
CUDA used to build PyTorch: 11.6
ROCM used to build PyTorch: N/A
OS: Microsoft Windows 10 Pro
GCC version: Could not collect
Clang version: Could not collect
CMake version: Could not collect
Libc version: N/A
Python version: 3.9.13 (tags/v3.9.13:6de2ca5, May 17 2022, 16:36:42) [MSC v.1929 64 bit (AMD64)] (64-bit runtime)
Python platform: Windows-10-10.0.19045-SP0
Is CUDA available: False
CUDA runtime version: 11.6.124
CUDA_MODULE_LOADING set to: N/A
GPU models and configuration: GPU 0: NVIDIA GeForce RTX 2060 SUPER
Nvidia driver version: 511.65
cuDNN version: Could not collect
HIP runtime version: N/A
MIOpen runtime version: N/A
Is XNNPACK available: True
Versions of relevant libraries:
[pip3] numpy==1.23.5
[pip3] torch==1.13.0+cu116
[pip3] torchaudio==0.13.0+cu116
[pip3] torchvision==0.14.0+cu116
[conda] Could not collect
323:デフォルトの名無しさん
22/11/26 12:10:47.07 8YfGLTST.net
pytorchはよく知らんけど
toolkit入れた？
cuda.is_availableてtorchのメッセージだよね？
てな感じで全然情報不足だわな
324:デフォルトの名無しさん
22/11/26 12:51:08.87 GN65Kd03.net
そうです。Torchのメッセージです。
CUDA Toolkitなら入れてますね。
現在は11.6.2をインストールしています。
以下コマンドプロンプトの出力結果。
■Nvcc -V で表示される内容
nvcc: NVIDIA (R) Cuda compiler driver
Copyright (c) 2005-2022 NVIDIA Corporation
Built on Tue_Mar__8_18:36:24_Pacific_Standard_Time_2022
Cuda compilation tools, release 11.6, V11.6.124
Build cuda_11.6.r11.6/compiler.31057947_0
■print(torch.__version__)で表示される内容
1.13.0+cu116
■torch.cuda.is_available()で表示される内容
False
過去にcuda_11.3.1とか10.2をツールキットからインストールした事もありますが、それでもダメでした。
325:デフォルトの名無しさん
22/11/26 12:53:48.55 8YfGLTST.net
まずcudaは動いてるの？
devicequeryとかsampleのプログラムが動くかどうか確認して問題なければ
torch周りの問題かと
326:デフォルトの名無しさん
22/11/26 22:03:13.46 GN65Kd03.net
ありがとうございます！
サンプル試したことなかったのでやってみました。
URLﾘﾝｸ(github.com)
上記サイトからCUDA Samples v11.6をダウンロードして、VisualStudio2017.slnをビルド
ビルド結果：ビルド: 175 正常終了、11 失敗、0 更新不要、0 スキップ（ただしビルド文書の中にいくつか失敗という文字あり）
deviceQueryをコマンドプロンプトから実行したところ・・・。
cudaGetDeviceCount returned 100
-> no CUDA-capable device is detected
Result = FAIL
という結果だったので、CUDA対応デバイスが検出されていないようです・・・。
327:デフォルトの名無しさん
22/11/26 22:06:59.94 GN65Kd03.net
あっ、ビルド結果、11個失敗しているって意味なんですね・・・。
328:デフォルトの名無しさん
22/11/26 23:52:01.01 HQ9nVBdA.net
GPUがちゃんと刺さって無いとか、電源不足とかドライバがちゃんとインストールされてないとか。
329:デフォルトの名無しさん
22/11/28 08:14:28.58 MM544pas.net
ん～、GPU自体は認識されてるみたいなんですよね。
GPU-Zで調べてみたらCUDAのところチェック外れてたので、やはりCUDAは認識されていないみたいです。
URLﾘﾝｸ(gpuz.techpowerup.com)
現在の電源は550Wなのですが、もうちょっと大きい容量のに変えてみて、それでダメならやはりドライバが合っていないという事でしょうか・・・。
330:デフォルトの名無しさん
22/11/28 08:33:47.92 D2ZVp0By.net
gpu-zはハードウェア情報なのでcudaのチェックが外れてるって
何かがかなりおかしい
20年前のgeforceなら分からんでもないがRTXなら
偽物をつかまされたか、ハードウェア故障を疑った方が良い気がする
331:デフォルトの名無しさん
22/11/28 08:52:47.95 FFe6eh0R.net
まずはsystem32にnvcuda.dll があるかどうか確認するかな
332:デフォルトの名無しさん
22/11/28 12:53:30.32 MM544pas.net
316です。
Sytem32にnvcuda.dll入ってました。
やはりハードウェアの故障でしょうか？ （今のGPUは約2年前にドスパラで買いました）
とりあえず電源ユニット注文しちゃったんで、電源交換＆GPUも付け直してみて、それでダメだったら新しいGPU購入ですかね(T_T)
333:デフォルトの名無しさん
22/12/02 20:33:45.27 2cIkFTzc.net
グラボを3060TIに変えてみましたが、それでもGPU-ZでCUDAにチェックが入っていませんでした・・・。
334:デフォルトの名無しさん
22/12/02 20:51:14.24 NwS5Fypu.net
あとはマザボかね～？？？
335:デフォルトの名無しさん
22/12/02 20:53:44.01 NwS5Fypu.net
差す穴変えてみた？
336:デフォルトの名無しさん
22/12/02 21:26:41.40 xlTggD6P.net
12V繋いでないってことはないよな
337:デフォルトの名無しさん
22/12/02 22:03:05.09 2B2OMrFT.net
>>332
Linuxでも入れて確認してみたら？
338:デフォルトの名無しさん
22/12/02 22:11:35.41 2cIkFTzc.net
接続するPCIEスロットも変えてみましたが・・・ダメでした。
グラボ上部にある補助電源？用のケーブルはちゃんと繋いでるはずなのですが、他に必要な接続とかあるんでしょうか？
現状のGPU-Zでの結果
URLﾘﾝｸ(gpuz.techpowerup.com)
339:デフォルトの名無しさん
22/12/03 06:08:21.14 9AvqKyUf.net
OpenGL, DirectX, vulkanとかのグラフィッ関係のAPIを使ったプログラムやゲームも動かないの?
本当にハードウェア関係に問題があるんならそういうプログラムも動かないだろうけど。
340:デフォルトの名無しさん
22/12/03 12:15:17.56 s9fm/abx.net
他のGPU全部殺してドライバアンインストール
3060Tiのドライバ入れなおせ
341:デフォルトの名無しさん
22/12/03 12:19:52.15 Xj+KmoE3.net
別のドライブにOSクリーンインストールしてみれば
342:デフォルトの名無しさん
22/12/04 11:39:28.03 2Pwe8xJ8.net
>>335
343:デフォルトの名無しさん
22/12/06 21:28:34.09 VA2y8qjn.net
316です。
思い切ってCドライブ初期化してみたらCUDA認識されました。
皆さんありがとうございました。
344:デフォルトの名無しさん
22/12/06 22:09:54.37 lBiSjHRF.net
経験上、一番難しいと感じたのはpython
listとかタプルとかデータ型が沢山あってこの変換がやたら発生
pandaとnumpyでも型が違うとか
torchは経験ないけどtensorflowは理屈が分かればそれほど難しくないと思う
345:デフォルトの名無しさん
22/12/07 02:07:47.36 imjH6UhZ.net
>>344
普通にC++の方が難しいと思うけど、どんな言語を試した経験上なの？
346:デフォルトの名無しさん
22/12/07 06:20:47.46 YHin935u.net
もしかしてCUDAでC++を覚えようとしてるの？
CUDAはCの知識でもできるけどCは？
347:デフォルトの名無しさん
22/12/07 13:28:40.27 wqy1K1SQ.net
C++は変態過ぎて
経験豊富ならべつにC++かまわんが
経験無いならCからやった方が良い
348:デフォルトの名無しさん
22/12/07 17:47:48.05 4gcyj4i9.net
>>347
Cは便利じゃないけど、確実だよね。マイコンとかでC++を使うのは、開発時間が十分に取れるなら(慣れてるなら)いいけど、処理系の変な制限にハマりそうで怖くて使えないわ。
349:デフォルトの名無しさん
22/12/23 20:34:26.37 ovzWFFv+.net
処理系の変な制限て何？
350:デフォルトの名無しさん
22/12/23 20:38:43.13 ovzWFFv+.net
小規模マイコンでもC++は便利だぞ
351:デフォルトの名無しさん
22/12/27 00:37:40.45 szDr1TAR.net
>>349
お気持ちC++実装でフル規格満たしてないとか普通にある。調べるのが手間すぎる。
352:デフォルトの名無しさん
22/12/27 04:54:17.83 Y8dS0LJm.net
フル規格を満たしてない事が普通にあるのはCも同じ
恐くて使えないのは経験が少ないからだな
353:デフォルトの名無しさん
22/12/27 13:10:12.33 vQ4ixuph.net
MSVCで作っておけば大体オッケー
g++依存は死ぬ
354:デフォルトの名無しさん
22/12/28 21:30:57.66 VGHawTJO.net
今時C++で開発してる会社あるのかな？
355:デフォルトの名無しさん
22/12/28 21:33:16.19 IIYChppW.net
当然
高速処理が要求されるようなところはC/C++だよ
356:デフォルトの名無しさん
23/01/14 11:12:05.67 hF49Vv2k.net
質問です。
cudaの9.0バージョンで
nvcc -V
このようなことをうっても、一切反応しないです。
誰か教えてえらいひとーー！
357:デフォルトの名無しさん
23/01/14 11:36:54.44 ITBYYn4q.net
＞＞一切反応しないです
もう少し具体的に
358:デフォルトの名無しさん
23/01/14 13:28:49.96 HwrORXYU.net
無修正を観ても、一切反応しないです。
誰か教えてえろいひとーー！
359:デフォルトの名無しさん
23/01/23 20:21:43.20 VfNttDi/.net
NVIDIA HPC SDK使ってる人いる？
windows版がないんだけどそのうちサポートされるんかね？
360:デフォルトの名無しさん
23/01/23 20:47:20.55 1CzktcoW.net
HPC SDKになった2020年にはWindows版は翌年公開予定と言っていたけど，
その記載もなくなったし永遠に出ない可能性が高いと思われる
当時とは違ってCUDA on WSL2でHPC SDKも使えるようになったから
Windows需要もそこで解消されるし
361:デフォルトの名無しさん
23/01/25 11:08:54.16 /YL2yMwg.net
>>360
詳しい人降臨ｷﾀｰ
旧PGIのユーザはほぼLinuxばっかだったようだしWindowsは見捨てられたのかと思ったけど必ずしもそうじゃないのかな
WSLで本当に性能出るの？ってのは気にはなるけど
ありがとうございました
362:デフォルトの名無しさん
23/03/05 20:30:51.25 skhIF3To.net
てst
363:デフォルトの名無しさん
23/03/05 21:06:34.13 skhIF3To.net
亀レス
>>354
あるっしょ
いくらでも,
このまえいびられて死んだ三菱電機の社員は電気系でも物性系の修士出てるのに
会社に入ってC++まともに書けなくて死んだ
物性選んだ時点でプログラム苦手なのにいきなりC売り物用のC++コード書け,
しかも,意地の悪い上司が適当なサンプルコードも見せてくれなきゃ死にたくなるかもな
364:デフォルトの名無しさん
23/05/14 05:25:29.03 EUIAlyu+.net
test
365:デフォルトの名無しさん
23/05/14 06:02:02.90 EUIAlyu+.net
おまえら教えろください
cudaDeviceSynchronizeが将来サポート打ち切りになることが決定し,
CUDA 12では既にオプションを指定しないとコンパイルできなくなっています.
これの代替APIは何でしょうか?
__global__関数内での待ち合わせ処理はどうすればいいのでしょう?
例
グローバルメモリに1M(1kx1k)の数値データがあります.そこから最大値を検索したい.
プログラムとしては
1Mデータを1kごとにグループ分割します.合計1k個のグループができます.
1つのグループを1つのスレッドに割り当て最大値を検索します. 当然スレッド数も1k個立ち上げます.
各スレッドは割り当てられたグループの1k個の中から最大値を検索し,見つけた値をシェアードメモリに書き込みます.
シェアードメモリもスレッド数と同じ1k個の配列から成リ, 各スレッドIDに紐づけされたアドレスに書き込みます.
例えばシェアードメモリを配列submaxとしたとき,
submax[threadIDx.x] = each_group_max;
みたいな感じです.
このとき各スレッドの書き込み完了を待ち合わせるのにcudaDeviceSynchronizeを使います.
書き込み完了後, 特定スレッドIDをもつスレッド(例えばthreadIDx.x==0)だけが1k個のsubmax中の最大値を検索することで
1M個データの最大値が決まります
これですとatomic関数を一切使わず高速処理が可能となります
366:デフォルトの名無しさん
23/05/14 06:21:14.34 EUIAlyu+.net
Dynamic parallelismでも使える実行時間測定関数でも
キャプチャ直前にcudaDeviceSynchronize();をコールしています
__device__
long long int
CaptureTimer() noexcept {
auto timer = (long long int)0; //コンパイラが文句いうので初期化しました
cudaDeviceSynchronize();
asm volatile("mov.u64 %0, %globaltimer;" : "=l"(timer));
return timer;
}
cudaStreamSynchronize()じゃ__global__でコールできないので駄目なんですよ
367:デフォルトの名無しさん
23/05/28 07:37:34.81 rAN5cbTU.net
何もレスないけど
cudaDeviceSynchronize なくても困らないコード書いてるの?
shared memory使うときどーしても使うと思うんだがこのAPI
368:デフォルトの名無しさん
23/05/30 17:45:24.83 GWh7Fu6m.net
ふつうは __syncthreads() 使うなあ。
というかカーネル内から全スレッド同期出来たとは知らなかった。
やっぱりいろいろ無理があったから廃止になったんじゃね？
369:デフォルトの名無しさん
23/06/01 19:34:35.47 woZHz0xe.net
カーネルの実行完了を待つ関数をカーネル内から呼び出せるのはおかしいだろ
370:デフォルトの名無しさん
23/06/02 21:14:34.84 GpIK8Zp9.net
>>368
>>369
お前らマジでレベル低いわ
シェアードメモリとかまったく使いこなせてないだろ
シェアードメモリってのはブロック内スレッドでのみ共通のメモリ
カーネル関数内で待ち合わせ処理できずにどうやって使うんだこんなもん
371:デフォルトの名無しさん
23/06/02 21:16:18.40 GpIK8Zp9.net
>>369にいたってはCUDA Dynamic Parallelizmすら理解してない
372:デフォルトの名無しさん
23/06/02 21:29:40.49 GpIK8Zp9.net
>>368
ふつうは?
__syncthreads()自体とっくに廃止されてそのかわりのcudaDeviceSynchronize()だろ
373:デフォルトの名無しさん
23/06/02 21:32:58.33 GpIK8Zp9.net
>>368
CUDA Runtime API rev.1 をみても __syncthreadsなんて載ってないんだが,
どこに載ってるか教えてくれ
374:デフォルトの名無しさん
23/06/02 22:14:21.99 GpIK8Zp9.net
>>368
URLﾘﾝｸ(docs.nvidia.com)
↑から最新のAPIマニュアルをダウソして検索したが __syncthreadsなんてものは存在しない
お前の書いたコードが最新のcuda 12.1.1コンパイラで使えるかどうか確認してみろ
とっくにサポートが打ち切られたAPIを使うと嘯吹いてただけじゃねーのか?
お前はとんでもなく長い間cudaコードを書いてなかったんじゃないのか?
375:デフォルトの名無しさん
23/06/02 23:26:41.73 zbQswG7U.net
カーネルからRuntime API使うのがそもそも違和感あるんだがイマドキのCUDAはそういうものなんか？
376:デフォルトの名無しさん
23/06/02 23:38:04.08 44SgMOSU.net
>>372
同期する範囲がぜんぜん違ぇだろうが
377:デフォルトの名無しさん
23/06/03 01:32:33.61 f1mb9frl.net
submit済みカーネルが2並列同じキューで動いていて、両方が完了待ちしたらデッドロック起こすなｗｗｗ
378:デフォルトの名無しさん
23/06/03 10:34:44.84 hgrjzBME.net
プログラムわかってないやつがいきなりの攻撃性を発揮するのガイジっぽくていいね
379:デフォルトの名無しさん
23/06/03 16:01:42.94 f1mb9frl.net
>>374
ランタイムAPIｗｗｗｗ
こっち見とけ
URLﾘﾝｸ(docs.nvidia.com)
380:デフォルトの名無しさん
23/06/10 08:16:47.35 gJM3u8Zc.net
cudaDeviceSynchronizeはこんなときに使う. Ki=1024, Mi = Ki*Ki, Gi = Ki*Miとでもして
__global__ void
cuda_main(){
double *idata = new [Mi]; double *odata1 = new [Mi]; double *odata2 = new [Mi];
body1<<< Ki, Ki >>> ( idata, odata1 ); //マルチスレッド実体
cudaDeviceSynchronize();
body2<<< Ki, Ki >>> ( odata1, odara2 ); //マルチスレッド実体
cudaDeviceSynchronize();
for( int i = 0; i < Mi; i++ ){
cudaDeviceSynchronize(); // *** こいつは毎回要ったと思う
printf(" %d %e\n", i, odata2[i]); //計算結果表示
}
cudaDeviceSynchronize(); // 上の***だけで こいつはなくても構わなかったと思う
delete[] odata2; delete[] odata1; delete[] idata;
}
main(){
//cuda 内newで確保するメモリが8MBを超える場合は設定要
cudaDeviceSetLimit(cudaLimitMallocHeapSize, size_t(Gi + Gi));
//printf fifoを16Miにしてみた
cudaDeviceSetLimit(cudaLimitPrintfFifoSize, 16 * Mi);
cuda_main<<<1、1>>>();
}
381:デフォルトの名無しさん
23/06/10 08:17:55.93 gJM3u8Zc.net
つづき
ホスト側からcuda用のエントリポイント cuda_main()を一つ起動して
そこで計算に必要なgpuメモリをnew/mallocで確保
delete/freeもしくはデストラクタでメモリ解放してほとんどふつーのC++プログラム作成
エントリポイントとなるcuda_mainスレッドを一つ起動するだけでほぼgpu側だけで閉じたふつーのc++コードになる
gpuが計算した結果をどー出力するか?ってのはあるんだが、
__global__関数内のprintfが標準出力にちゃーんと出力してくれる
つまり、 リダイレクトでファイルに落とせる
コマンドラインから
> cuda_program.exe >> file.dat
て感じ
382:デフォルトの名無しさん
23/06/10 08:20:34.56 gJM3u8Zc.net
つづき
こういうコードスタイルにすることでcudamallocで確保するgpuメモリはホスト<->gpuのインターフェースに使う最低限のメモリに限定することができて
コーディングが格段に楽になる ほぼふつーのC++プログラムと変わらん
この書き方で問題なのは__global__内でunique_ptrが使えないことなんだが、
数値計算でunique_ptr使えなくて困る場合ってどんな場合か逆に聞きたいわ.
それより
ホスト側でgpu内部処理にしか使わないメモリまで確保して、そいつをcuda::unique_ptrなんか作って管理するよりよほど楽だと思うが?
こういうのはディレクトリの外からディレクトリ操作するのに似た感覚でとても耐えられんし、
cuda Dynamic Parallelismと__global__関数内でnew/deleteを書けるようになってるお陰でふつーのC++コードが書けるようになってるんだわ.
んで肝心の cudaDeviceSynchronize() だが
マルチスレッド実態は body1、 body2だがbody2の実行と計算結果odata2の表示はbody1とbody2の計算完了が前提としてる
それらの待ち合わせ処理としてを使うんだが、
ここで cudaDeviceSynchronize() がないと dynamic Parallelismがあっても上のようなコードは書けない
383:デフォルトの名無しさん
23/06/10 10:12:15.01 5lxShvGN.net
>>380
同じstream使ってんのにいちいちそんなところで同期してたら効率悪くね？
384:デフォルトの名無しさん
23/06/10 11:53:19.12 Y9DL3tjK.net
まだやってたのかよw>>365
385:デフォルトの名無しさん
23/06/16 10:37:03.14 VMczRTMU.net
>>383
いちいちも何も
body1<<< Ki, Ki >>> ( idata, odata1 ); //マルチスレッド実体
cudaDeviceSynchronize();
body2<<< Ki, Ki >>> ( odata1, odara2 ); //マルチスレッド実体
cudaDeviceSynchronize();
body2の処理開始はbody1の計算終了が前提としたプログラムのつもりで書いてるのでこの場合は必ず必要なわけで
この2つのの同期は
マルチスレッド起動箇所がホスト側ならホストで
マルチスレッド起動箇所がgpu側ならgpuでやることが必須なんだが
gpuの側でやる意味はふつーに_global__関数の中で, ローカルメモリの管理にnew/delete使えること
cpuのアクセラレータとしてGPUを使うんじゃなくGPUだけで完結するならこう書くでしょーよ
わざわざcpu側に制御戻す必要なく処理速度が落ちるわけじゃない.
それどころかはるかにプログラムの見通しもよくなるんだから
386:デフォルトの名無しさん
23/06/16 20:00:03.41 duxPCvpi.net
丸々一ヶ月経ってもまだ躓いてるのかw
せっかくアドバイス貰っても理解する気がないならしょうがないな。
387:デフォルトの名無しさん
23/06/16 23:41:50.06 T6+41XgS.net
>>385
dynamic parallelismがどうたらイキってた輩が実はstreamすらろくに理解していなかったという
388:デフォルトの名無しさん
23/06/17 09:32:02.85 6mG7lpSl.net
そもそもShared Memoryが云々の話とこれ全く関係ないし、unique_ptrがどうとか誰も聞いてないし、ただイキってクソコード書いてるnoobにしか見えん
389:デフォルトの名無しさん
23/06/30 06:45:59.93 HGq0NS3h.net
>>387
streamてCPUとGPUの協調のこと言ってるの?
協調はこの際,まったく関係ないんだが,
CPUからGPUに制御を移して完全にCPUとは分離する話をしてんの
そのほうが遥かにプログラムが楽に簡潔に書けるからそれを言ってる.
GPUのほうが8倍から1桁速度が速いので実験的な計算ならこれで十分な速度が得られる
CPUとの協調とか難しいプログラムを書くんじゃなく
GPUだけで計算を完結させるためにDynamic Parallelismを使う
Dynamicに並列化したいわけじゃなくGPUのプロセスを一つ立ち上げたいためにそれを使う
390:デフォルトの名無しさん
23/06/30 06:53:13.13 HGq0NS3h.net
>>386
何のアドバイス?ww
GPUプロセスの中で全プロセスどうやってwaitするのさw
ま, cudaのユーザサイトには代替案とか書いてる人いるよ
cudaDeviceSynchronize()と比較して速度は期待できないが一応使える.
せめてそれぐらいのアドバイスできるようになってから言いな.
俺のようなコード書いてGPUプロセスの中からcudaDeviceSynchronize()コールしてるユーザは一定数いるってこった
どーせ__global__でnew/deleteふつーに使えること知らんかったんだろお前ww
391:デフォルトの名無しさん
23/06/30 07:16:01.83 HGq0NS3h.net
>>388
誰も聞いてないunique_ptrて
CPU側でGPUのマルチスレッド実体を起動しようとすると,
GPU内部でしか利用しないローカルメモリまでCPUから管理する羽目になる
そのときcudamalloc/cudafreeなんか使ってたんじゃメモリ管理が大変で,
gpu::unique_ptrでも作らないとCコード書いてるのと変わらんことになるから言ってる
そんなことせずとも__global__でnew/deleteがデフォで使えるんだからメモリ管理はC++03並には書ける
__global__でnew/deleteを使ってGPUだけで完結処理するには
親スレッドの__global__内で子スレッドの待ち合わせ処理がどうしても必要になる場合があるから言ってる
できるもんなら>>380のcudaDeviceSynchronize(); を同等処理に置き換えてみ
あ, すでに cudaのユーザーサイトに投稿してる人が居るんで探せばしまいだけどねww
392:デフォルトの名無しさん
23/06/30 08:24:20.23 pBAX8tCp.net
俺はCUDA使ったこと無いんだけど
＞__global__でnew/deleteを使ってGPUだけで完結処理
これって思い込みじゃねえかなあ
393:デフォルトの名無しさん
23/06/30 11:03:40.92 HGq0NS3h.net
>>392
作ったことないんだろww
こっちは書いてるから言ってる.
>>380を見ろ
この構成の場合がそれに該当する
CPUはgpu内でnew/deleteに使う最大値のみ設定する.デフォルトは小さいんでな
例えば
1.GPUで用意してる数学関数だけでできる数値計算
2.信号もノイズも乱数から作るモンテカルロ・シミュレーション
要はファイルやなんらかのデバイス,ネットから読み込んでその処理をGPUに転送する場合は
CPUは無関係じゃいられないが, 上のような場合はGPUだけで計算可能だ.
具体例として
1. f(x) = ∫_0^x sin(t)/t dt 0<= x <= 10 を計算して計算結果をファイルにセーブする
2. ある変調方式のエラーレートをシミュレーションで計算し. ファイルにセーブする.
これらならCPUとほぼ無関係に計算できる ファイルや外部デバイスからデータを読み込んでくる必要がないんでね.
あと__global__でのprintfはふつーにリダイレクトが有効. GPU内でファイルオープンとかの必要もなくファイルに落とせる
ま,これらは極端な例だが, 最低限, 初期値とか処理データ, GPUで計算終了後の
GPU-CPU間のインターフェース用メモリのみcudamalloc/dudafreeで管理してやればよくて
GPU内部だけで使うようなメモリはCPU側で一切管理する必要がなくなる.
gpu内ではC++03レベルのnew/deleteしか使えないけどな.
394:デフォルトの名無しさん
23/06/30 11:07:50.35 HGq0NS3h.net
>>392
ウソだと思うなら,>>380のコードを実際cudaでコンパイルして確認してみろや
実際__global__内でnew/malloc使えて__global__から__global__を起動できることだけがポイントなんだから
ちなみに, dynamic parallelismサポートしてないような古いグラボは使用禁止な.
395:デフォルトの名無しさん
23/06/30 11:38:34.15 HGq0NS3h.net
>>392
new/deleteが__global__内でふつーに使えるってことなら
cuda c programming guide 最新版の
7.34. Dynamic Global Memory Allocation and Operations
どっちが思い込みかね?ww
396:デフォルトの名無しさん
23/06/30 18:28:44.41 h/tbEWPQ.net
もしかしてNVIDIAのフォーラムでモデレータに I’m confused とか言われてんのコイツ？
397:デフォルトの名無しさん
23/06/30 22:03:38.67 nvcNe2IT.net
炎上学習法かとも思ったが全然学習してる素振りもないのよなあ
>streamてCPUとGPUの協調のこと言ってるの?
ちょっとはググるなりしたらどうかね
> body1<<< Ki, Ki >>> ( idata, odata1 ); //マルチスレッド実体
> cudaDeviceSynchronize();
> body2<<< Ki, Ki >>> ( odata1, odara2 ); //マルチスレッド実体
body1とbody2は同じstreamだから基本的にそんなところにsynchronizeはいらんのよ
398:デフォルトの名無しさん
23/06/30 22:07:49.01 nvcNe2IT.net
>そのときcudamalloc/cudafreeなんか使ってたんじゃメモリ管理が大変で,
これも意味不明だなあ
malloc/freeと比べてどう大変だと言うんだろう
399:デフォルトの名無しさん
23/06/30 23:46:28.63 lTu2cFop.net
病気の人に関わってもいいことないですよ
400:デフォルトの名無しさん
23/07/02 11:33:55.45 ZzvN3CPt.net
同期オブジェクトが出てないんだから、そりゃシリアルに動くやろうなあ
CUDA知らん
401:デフォルトの名無しさん
23/07/02 14:47:53.06 tUgy2gS2.net
>>396
ゲラゲラ
それどこだよww
お前のプロファイル推定正しいかどうか見てから発表してやるよ
さらせよソコをよww
>>397
12.2 プログラムガイド pp47
For code that is compiled using the --default-stream legacy compilation flag, the default stream
is a special stream called the NULL stream and each device has a single NULL stream used for all
host threads. The NULL stream is special as it causes implicit synchronization as described in Implicit
Synchronization.
For code that is compiled without specifying a --default-stream compilation flag,
--default-stream legacy is assumed as the default.
11.4以降 --default-streamは非推奨. 当然このオプションはデフォルトでなくなり,
暗黙の同期ストリームであるNull streamはデフォルトではなくなった.
つまり, ストリームは何も指定しなければ非同期ストリームとなった
同期ストリームとなることを保証したい場合, 当然cudaStreamSynchronize()で挟むだろうが
そのすぐ下の6.2.8.5.3 Explicit Synchronizationも読んどけよ
コロコロ変わるデフォのコンパイルオプションに頼るお前
>>398
freeが毎回書いてられないといってるわけ. それとcudafreeが同じなのは当たり前な
だからホスト側ではgpu用のgpu::unique_ptrとかユーザ側でこさえてるんだろが.
これを書いてるサイトはいくらでもある.
__global__内ではこういうmake_uniqueは動かないのでc++11レベルでは書けないが,
普通にnew/deleteやC++03レベルのデストラクタが動作するので,
ホスト側でgpuのローカルメモリの解放を手でやらかすよりよほど楽だと言ってるのだよ.
402:デフォルトの名無しさん
23/07/02 19:20:21.21 nj7sKJew.net
>>401
いや、それストリーム間の同期の話であって同じストリームに投入したカーネルの実行の話とは関係ないんだが。
いろいろ検索したりしたんだろうけどここ勘違いするようなレベルでイキりまくられても。
もともとは cudaDeviceSynchronize がデバイス上で非推奨になるんで代替策をどうするかって話だったと思うけど、
「自分は間違ってない」と強弁するだけならそりゃ2ヶ月近く解決しないわけだわ。
403:デフォルトの名無しさん
23/10/20 12:42:50.23 /M3RKJCH.net
NVIDIA中國から撤退
404:デフォルトの名無しさん
23/12/07 06:26:05.21 R1AO2r8W.net
PCくそど初心者で、オーディオをやっている者です。
現在PC（linux）でHQplayerというソフトで音楽ききてます。
そのPCにグラボ刺したら音質上がるなんてガセネタつかまされたんで、騙されてみようと思うのですが、その際CUDAの設定しないとGPUが使われないということのようなのです。
調べたら、ドライバー、ツールキット、cuDNNの３つをインストールするみたいなのですが、最後のcuDNNの意味がわかりません。
有識者の方、どうかご教授下さい！
405:デフォルトの名無しさん
23/12/10 12:43:06.09 9z8kD1aM.net
>>404
他人の作った何をやっているかわからないプログラムを使うより
音とプログラミングとCUDAについて勉強して
自分で音質を上げるプログラムを作ったほうがよいと思うよ。
本屋とかアマゾンでそういう専門書も探せば見つかるでしょ。
406:デフォルトの名無しさん
23/12/13 06:47:58.97 YKWD9gfa.net
>>405
398です。
なるほど、そういう方法もあるのですね。
自分で勉強してプログラムのことまで理解するなんて大変そうですが、根本を理解していないと使う意味ないですもんね。
大変そうですが、チャレンジしてみます。
ありがとうございました。
407:デフォルトの名無しさん
24/01/13 07:53:38.21 Ce28ohDx.net
RTX4090よりA4000をおすすめしてる所がありますが、これは長期稼働の安定性と低消費電力が理由でしょうか。
lstmで出来るだけ早く学習させたいのですが4090の方が早そうですがいかがですか。
408:デフォルトの名無しさん
24/03/14 19:33:05.60 LpPl9eeZ.net
自己解決
webページ作成者に聞いた
4090の方が速いが、例えば3時間が4時間になったところでたいした違いはないでしょうとか
いくつものパターンを試したいので少しでも早いものをってことで4090を買った