[[ノート>ノート/ノート]]~
訪問者 &counter();  最終更新 &lastmod();~
#contents

旧バージョン [[ノート/CUDA/CUDA設定onFedora12]]

**Titanを入手したので、Fedora19上にCUDA5.5をインストールする(2013-07-08) [#h8f0546b]

***OSインストール [#d23825e3]
Centos6.4はまだマザーボードのチップセットに対応していなかったので、Fedora 19にする。

-Live CDをUSBにコピー([[LiLi USB Creatorを使用:http://www.linuxliveusb.com/]])
-USBから起動
-HDDへインストール
-起動後、yum updateで更新

***CUDA 5.5 インストール (2013-07-08) [#lbaea441]
-[[参考サイト :http://hobiger.org/blog/2013/07/04/fedora-19-and-cuda/]]

-現時点でのLinux (64) 用のドライバーを[[ダウンロード:http://www.geforce.com/drivers/results/63577#]]
 /etc/default/grub の中で、GRUB_CMDLINE_LINUX= の末尾にmodeset=0を追加して
 grub2-mkconfig -o /boot/grub2/grub.cfg
 再起動
 yum install gcc kernel-devel
 init 3                 <<< Xサーバーを止めるのに必須
 init 3                 <<< Xサーバーを止めるのに必須(XWinに戻すにはinit 5)
 持ってきたドライバ NVIDIA....run を実行 ファイルモードを755にして ....run -Z
-CUDA5.5の環境を[[ダウンロード:https://developer.nvidia.com/cuda-pre-production?sid=306057]] ここでダウンロードされるrpm(cuda-repo-fedora18-5.5.0.x86_64.rpm)を実行(rpm -i ファイル名)すると、レポジトリ情報(/etc/yum.repos.d/cuda.repo)への書込みファイルがダウンロードされるだけ。
 rpm -i cuda-repo-fedora18-5.5.0....rpm
-実際のcuda環境のダウンロードは、このレポジトリ情報が作られたことを踏まえて、yum install cuda でダウンロード・インストール
 yum install cuda    <<< これはやらないほうが良いらしい
-[[使い方ドキュメントはここ:http://developer.download.nvidia.com/compute/cuda/5_5/rc/docs/CUDA_Getting_Started_Linux.pdf]]
-CUDAコンパイルしてみたらlsb_releaseコマンドが無いと言われたので、lsbをインストール
 yum install lsb

***一般ユーザが、自分のホームディレクトリの下へCUDA5.5開発環境をインストールするには [#b92b1a7c]

システムレベルのインストールだと、/usr/local/cuda-5.5の下に開発環境(samples)を展開し、一般ユーザにとってはread-onlyになってしまう。

一般ユーザが開発できるためには、コマンド
 cuda-install-samples-5.5.sh ./      (但し「./」はインストールしたい場所)
を実行する。そうすると
 Copying samples to .//NVIDIA_CUDA-5.5_Samples now...
 Finished copying samples.
で、今いる下「./」に NVIDIA_CUDA-5.5_Samples というディレクトリを作って、その中にサンプルファイルが(書き込み可能で)コピーされる。
ここを使って、プログラムを作ればよい。

***Samples がコンパイルできるようになった [#u6da8614]
以下の話は、/usr/local/cuda-5.5においた(システム用の)開発ディレクトリ上で実験している。一般ユーザは、自分のディレクトリの下に展開したNVIDIA_CUDA-5.5_Samplesの下であると読み替えてほしい。

-/usr/local/cuda-5.5/samples/ にあるサンプルをコンパイルする
 cd /usr/local/cuda-5.5/samples
 make
これでmakeが終わると、バイナリは/usr/local/cuda-5.5/samples/bin/x86_64/linux/releaseにできるので
 cd /usr/local/cuda-5.5/samples/bin/x86_64/linux/release

-Samplesの中の deviceQuery を実行してみる
 ./deviceQuery
 ./deviceQuery Starting...
 
  CUDA Device Query (Runtime API) version (CUDART static linking)
 
 Detected 1 CUDA Capable device(s)
 
 Device 0: "GeForce GTX TITAN"
   CUDA Driver Version / Runtime Version          5.5 / 5.5
   CUDA Capability Major/Minor version number:    3.5
   Total amount of global memory:                 6143 MBytes (6441730048 bytes)
   (14) Multiprocessors x (192) CUDA Cores/MP:    2688 CUDA Cores
   GPU Clock rate:                                876 MHz (0.88 GHz)
   Memory Clock rate:                             3004 Mhz
   Memory Bus Width:                              384-bit
   L2 Cache Size:                                 1572864 bytes
   Max Texture Dimension Size (x,y,z)             1D=(65536), 2D=(65536,65536), 3D=(4096,4096,4096)
   Max Layered Texture Size (dim) x layers        1D=(16384) x 2048, 2D=(16384,16384) x 2048
   Total amount of constant memory:               65536 bytes
   Total amount of shared memory per block:       49152 bytes
   Total number of registers available per block: 65536
   Warp size:                                     32
   Maximum number of threads per multiprocessor:  2048
   Maximum number of threads per block:           1024
   Maximum sizes of each dimension of a block:    1024 x 1024 x 64
   Maximum sizes of each dimension of a grid:     2147483647 x 65535 x 65535
   Maximum memory pitch:                          2147483647 bytes
   Texture alignment:                             512 bytes
   Concurrent copy and kernel execution:          Yes with 1 copy engine(s)
   Run time limit on kernels:                     Yes
   Integrated GPU sharing Host Memory:            No
   Support host page-locked memory mapping:       Yes
   Alignment requirement for Surfaces:            Yes
   Device has ECC support:                        Disabled
   Device supports Unified Addressing (UVA):      Yes
   Device PCI Bus ID / PCI location ID:           1 / 0
   Compute Mode:
      < Default (multiple host threads can use ::cudaSetDevice() with device simultaneously) >
 
 deviceQuery, CUDA Driver = CUDART, CUDA Driver Version = 5.5, CUDA Runtime Version = 5.5, NumDevs = 1, Device0 = GeForce GTX TITAN
 Result = PASS
となった。

&ref(2013-07-08_TitanSpec.png);~
(www.hardwarecanucks.com)

**gpupiを実行してみる(2013-07-09) [#i452db0e]
***CUDAの変更点 [#x60ea359]
-#include <cutil.h>がなくなる~
代わりに
 // CUDA runtime
 #include <cuda_runtime.h>
 // helper functions and utilities to work with CUDA
 #include <helper_functions.h>
 #include <helper_cuda.h>
を追加
-同じように、初期化の部分で
 CUT_DEVICE_INIT(argc, argv);
がなくなって、代わりに
 int devID = findCudaDevice(argc, (const char **)argv);
を使うらしい。
-タイマ回りが変わっているようだ~
sampleのtemplateにある方法は
        StopWatchInterface *timer = 0;
        sdkCreateTimer(&timer);
 
        sdkStartTimer(&timer);
        何かする
        sdkStopTimer(&timer);
        printf("計算時間 =%f(ms)\n", sdkGetTimerValue(&timer));
        sdkDeleteTimer(&timer);
としている。~
~
NVIDIAのマニュアルでは~
5.1.2. Using CUDA GPU Timers~
The CUDA event API provides calls that create and destroy events, record events  (via timestamp), and convert timestamp differences into a floating-point value in milliseconds. How to time code using CUDA events illustrates their use.~
~
How to time code using CUDA events~
 cudaEvent_t start, stop;
 float time;
 
 cudaEventCreate(&start);
 cudaEventCreate(&stop);
 
 cudaEventRecord( start, 0 );
 kernel<<<grid,threads>>> ( d_odata, d_idata, size_x, size_y, 
                            NUM_REPS);
 cudaEventRecord( stop, 0 );
 cudaEventSynchronize( stop );
 
 cudaEventElapsedTime( &time, start, stop );
 cudaEventDestroy( start );
 cudaEventDestroy( stop );
Here cudaEventRecord() is used to place the start and stop events into the default stream, stream 0. The device will record a timestamp for the event when it reaches that event in the stream. The cudaEventElapsedTime() function returns the time elapsed between the recording of the start and stop events. This value is expressed in milliseconds and has a resolution of approximately half a microsecond. Like the other calls in this listing, their specific operation, parameters, and return values are described in the CUDA Toolkit Reference Manual. Note that the timings are measured on the GPU clock, so the timing resolution is operating-system-independent.~
~
としている。~

あとはほとんどそのままで動いた。変更後のプログラムは
 #include <stdio.h>
 //include <cutil.h>  Disappeared
 
 // CUDA runtime
 #include <cuda_runtime.h>
 
 // helper functions and utilities to work with CUDA
 #include <helper_functions.h>
 #include <helper_cuda.h>
 
 //#define FLOAT float
 #define FLOAT double
 
 #define BLOCK 2048    /* Number of blocks */
 #define THREAD 32  /* Number of threads per block */
 
 #define STRIPS 65536*256  /* Number of total strips */
 //#define STRIPS 65536  /* Number of total strips */
 
 void Host(FLOAT *Result);
 __global__ void Kernel(FLOAT *Result);
 
 FLOAT h_result[THREAD*BLOCK];
 
 int main(int argc, char** argv){
         StopWatchInterface *timer = 0;
         sdkCreateTimer(&timer);
         //unsigned int timer;
         FLOAT pi;
 
         //CUT_DEVICE_INIT(argc, argv);
         int devID = findCudaDevice(argc, (const char **)argv);
 
         FLOAT *d_result;
 
         cudaMalloc((void**) &d_result, sizeof(FLOAT)*THREAD*BLOCK);
         cudaMemset(d_result, 0, sizeof(FLOAT)*THREAD*BLOCK);
 
         sdkStartTimer(&timer);
 
         cudaMemcpy(d_result, h_result, sizeof(FLOAT)*THREAD*BLOCK, cudaMemcpyHostToDevice);
 
         fprintf(stderr, "THREAD=%d, BLOCK=%d\n", THREAD, BLOCK);
         dim3 grid(BLOCK, 1, 1);
         dim3 threads(THREAD, 1, 1);
         Kernel<<< grid, threads >>>(d_result);
 
         cudaMemcpy(h_result, d_result, sizeof(FLOAT)*THREAD*BLOCK, cudaMemcpyDeviceToHost);
 
         sdkStopTimer(&timer);
         printf("計算時間 =%f(ms)\n", sdkGetTimerValue(&timer));
         sdkDeleteTimer(&timer);
 
         pi=0.0;
         for (int k=0; k<THREAD*BLOCK; k++) {
                 pi += h_result[k];
                 //printf("Result[%d]=%15.12f\n", k, h_result[k]);
         }
         printf("GPU計算結果 =%15.12f\n",  (pi - (((FLOAT)0.5)/((FLOAT)STRIPS))) * (FLOAT)4.0);
 
         cudaFree(d_result);
 
         sdkCreateTimer(&timer);
         sdkStartTimer(&timer);
         Host(h_result);
         sdkStopTimer(&timer);
         printf("ホストの計算時間 =%f(ms)\n", sdkGetTimerValue(&timer));
         sdkDeleteTimer(&timer);
 
         pi=0.0;
         for (int k=0; k<THREAD*BLOCK; k++) {
                 pi += h_result[k];
                 //printf("Result[%d]=%15.12f\n", k, h_result[k]);
         }
         printf("host計算結果 =%15.12f\n",  (pi - (((FLOAT)0.5)/((FLOAT)STRIPS))) * (FLOAT)4.0);
         //getch();
 }
 
 __global__ void Kernel(FLOAT *Result)
 {
         //GPUでの処理
         int ix=blockIdx.x*blockDim.x + threadIdx.x;
         FLOAT tmp=0.0;
 
         for(int k=0; k<(STRIPS/BLOCK/THREAD); k++){
                 FLOAT x = ((FLOAT) ((ix*(STRIPS/BLOCK/THREAD))+k)) / ((FLOAT)STRIPS);
                 tmp+=((FLOAT)sqrt(1.0-x*x)) / ((FLOAT)STRIPS);
         }
 
         Result[ix]=tmp;
 }
 
 void Host(FLOAT *Result){
         //ホスト側での計算
         int k, ix;
         FLOAT x, tmp;
         for(ix=0; ix<(THREAD*BLOCK); ix++){
                 tmp = 0.0;
                 for(k=0; k<(STRIPS/BLOCK/THREAD); k++){
                         x = ( (((FLOAT)ix)*((FLOAT)(STRIPS/BLOCK/THREAD))) + (FLOAT)k) / ((FLOAT)STRIPS);
                         tmp=tmp+(((FLOAT)sqrt(1-(x*x))) / ((FLOAT)STRIPS) );
                         //printf("xi=%15.12f\n", ((FLOAT)sqrt(1.0-x*x)) / ((FLOAT)STRIPS));
                 }
 
                 Result[ix]=tmp;
 
         }
 }


実行結果は
 GPU Device 0: "GeForce GTX TITAN" with compute capability 3.5
 
 THREAD=32, BLOCK=2048
 計算時間 =3.423000(ms)
 GPU計算結果 = 3.141592653573
 ホストの計算時間 =359.904999(ms)
 host計算結果 = 3.141592653573
となった。

**NVIDIAのプロファイラ [#d2ace3e5]
titanでは/usr/local/cuda/bin/の下に、プロファイラが2つ入っている。
-nvvp  NVidia Visual Profiler~
使い方 ⇒ マニュアルだと[[ここ:http://docs.nvidia.com/cuda/profiler-users-guide/index.html]]([[ローカルコピーだとここ:http://titan.yy.is.sci.toho-u.ac.jp/cuda55/profiler-users-guide/index.html]]) とか [[fixsterのページ:http://gpu.fixstars.com/index.php/CUDA_Visual_Profiler%E3%82%92%E4%BD%BF%E3%81%86#Linux.E3.81.AE.E8.BF.BD.E5.8A.A0.E8.A8.AD.E5.AE.9A.E9.A0.85.E7.9B.AE]]

-nvprof  NVidia (コマンドラインの) Profiler~
使い方 ⇒ [[たとえばここ:http://topsecret.hpc.co.jp/wiki/index.php/CUDA_5%E3%81%AE%E6%96%B0%E6%A9%9F%E8%83%BD(4):_nvprof%E3%83%97%E3%83%AD%E3%83%95%E3%82%A1%E3%82%A4%E3%83%A9]] とか [[ここ:http://d.hatena.ne.jp/w_o/20121211]]

トップ   編集 差分 バックアップ 添付 複製 名前変更 リロード   新規 一覧 単語検索 最終更新   ヘルプ   最終更新のRSS