FPGA開発日記

FPGAというより、コンピュータアーキテクチャかもね! カテゴリ別記事インデックス https://sites.google.com/site/fpgadevelopindex/

Nsight with Visual Studio ファーストインプレッション

GPGPUデバッグ、プラファイラって無いの?ということでいろいろ探していたのだが、nVIDIAの配布するGPUSDKにはNsightというツールが入っているらしい。

とりあえずチュートリアルだけ使ってみたので、そのやり方をメモしておく。

  1. Visual Studio 2013 を立ち上げる。

f:id:msyksphinz:20150911002128j:plain

  1. 「新規プロジェクト作成」の中に、[Template]->[CUDA 6.5] というのが入っているのが分かる。これを使おう。

f:id:msyksphinz:20150911002252j:plain

  1. プロジェクトを作成すると、警告メッセージのようなものが出てくるが、気にせず進める。ここでは[Connect unsecurely]を選択した。

f:id:msyksphinz:20150911002439j:plain

  1. プロジェクトが開く。[Application Control]などのボタンが表示されている。とりあえずここまででテンプレートのプログラムが入っており、これだけでビルド可能だ。

f:id:msyksphinz:20150911002623j:plain

__global__ void addKernel(int *c, const int *a, const int *b)
{
    int i = threadIdx.x;
    c[i] = a[i] + b[i];
}

...

// Helper function for using CUDA to add vectors in parallel.
cudaError_t addWithCuda(int *c, const int *a, const int *b, unsigned int size)
{
    int *dev_a = 0;
    int *dev_b = 0;
    int *dev_c = 0;
    cudaError_t cudaStatus;

    // Choose which GPU to run on, change this on a multi-GPU system.
    cudaStatus = cudaSetDevice(0);
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaSetDevice failed!  Do you have a CUDA-capable GPU installed?");
        goto Error;
    }

    // Allocate GPU buffers for three vectors (two input, one output)    .
    cudaStatus = cudaMalloc((void**)&dev_c, size * sizeof(int));
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaMalloc failed!");
        goto Error;
    }

    cudaStatus = cudaMalloc((void**)&dev_a, size * sizeof(int));
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaMalloc failed!");
        goto Error;
    }

    cudaStatus = cudaMalloc((void**)&dev_b, size * sizeof(int));
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaMalloc failed!");
        goto Error;
    }

    // Copy input vectors from host memory to GPU buffers.
    cudaStatus = cudaMemcpy(dev_a, a, size * sizeof(int), cudaMemcpyHostToDevice);
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaMemcpy failed!");
        goto Error;
    }

    cudaStatus = cudaMemcpy(dev_b, b, size * sizeof(int), cudaMemcpyHostToDevice);
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaMemcpy failed!");
        goto Error;
    }

    // Launch a kernel on the GPU with one thread for each element.
    addKernel<<<1, size>>>(dev_c, dev_a, dev_b);

    // Check for any errors launching the kernel
    cudaStatus = cudaGetLastError();
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "addKernel launch failed: %s\n", cudaGetErrorString(cudaStatus));
        goto Error;
    }
    
    // cudaDeviceSynchronize waits for the kernel to finish, and returns
    // any errors encountered during the launch.
    cudaStatus = cudaDeviceSynchronize();
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaDeviceSynchronize returned error code %d after launching addKernel!\n", cudaStatus);
        goto Error;
    }

    // Copy output vector from GPU buffer to host memory.
    cudaStatus = cudaMemcpy(c, dev_c, size * sizeof(int), cudaMemcpyDeviceToHost);
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaMemcpy failed!");
        goto Error;
    }

Error:
    cudaFree(dev_c);
    cudaFree(dev_a);
    cudaFree(dev_b);
    
    return cudaStatus;
}

  1. [Trace Setting]から、"System" と "CUDA"を選んでおく。こうすることで、CUDAの動作状況と、システムで何が呼ばれたかをトレースすることができる。

f:id:msyksphinz:20150911002855j:plain

  1. でっかい赤ボタン[Application Control]の[Launch]をクリックする。ビルドと実行が開始される。実行が終了すると、統計情報が表示される。

f:id:msyksphinz:20150911003041j:plain

  1. ドロップダウンから[TimeLine]を選択すると、関数の呼び出しのタイムラインが表示される。

f:id:msyksphinz:20150911003155j:plain

  1. 拡大していくと、どのような関数が実行されているかが分かる。これだと、cudaMemCpyがかなりの時間を食っているなあ。。。

f:id:msyksphinz:20150911003236j:plain

これを使うと、今まで開発したCUDAのアプリケーションの性能をうまく測定できそうだ。いろいろ試してみよう。