concurrency - Tesla K20 を共有する 2 つのプロセスの同時実行

Question

GPU リソースを共有しながら同時に実行するためにカーネルの 2 つのインスタンスを起動すると、奇妙な動作が発生します。

スレッドが (ループを使用して) 操作を数回実行する単一の SM (マルチプロセッサ) で実行することを目的とした CUDA カーネルを開発しました。

カーネルはブロックのみを作成する準備ができているため、1 つの SM のみを使用します。

シンプル.cu

#include <cuda_runtime.h>
#include <stdlib.h>
#include <stdio.h>
#include <helper_cuda.h>
using namespace std;

__global__ void increment(float *in, float *out)
{
    int it=0, i = blockIdx.x * blockDim.x + threadIdx.x;
    float a=0.8525852f;

    for(it=0; it<99999999; it++)
             out[i] += (in[i]+a)*a-(in[i]+a);
}

int main( int argc, char* argv[])
{
    int i;
    int nBlocks = 1;
    int threadsPerBlock = 1024;
    float *A, *d_A, *d_B, *B;
    size_t size=1024*13;

    A = (float *) malloc(size * sizeof(float));
    B = (float *) malloc(size * sizeof(float));

    for(i=0;i<size;i++){
            A[i]=0.74;
            B[i]=0.36;
    }

    cudaMalloc((void **) &d_A, size * sizeof(float));
    cudaMalloc((void **) &d_B, size * sizeof(float));

    cudaMemcpy(d_A, A, size, cudaMemcpyHostToDevice);
    cudaMemcpy(d_B, B, size, cudaMemcpyHostToDevice);

    increment<<<nBlocks,threadsPerBlock>>>(d_A, d_B);

    cudaDeviceSynchronize();

    cudaMemcpy(B, d_B, size, cudaMemcpyDeviceToHost);

    free(A);
    free(B);

    cudaFree(d_A);
    cudaFree(d_B);

    cudaDeviceReset();

    return (0);
}

したがって、カーネルを実行すると:

time ./simple

私は得る

real 0m36.659s user 0m4.033s sys 0m1.124s

それ以外の場合、2 つのインスタンスを実行すると:

time ./simple & time ./simple

プロセスごとに取得します：

real 1m12.417s user 0m29.494s sys 0m42.721s

real 1m12.440s user 0m36.387s sys 0m8.820s

私の知る限り、実行は同時に実行され、1 つ (約 36 秒) 続く必要があります。ただし、それらは基本時間の 2 倍続きます。GPU には 13 の SM があり、それぞれが 1 つのブロックを実行する必要があることがわかっているため、カーネルは 1 つのブロックのみを作成します。

それらは同じ SM で実行されていますか?

異なる SM で同時に実行するべきではありませんか?

編集済み

わかりやすくするために、nvprof から取得した同時実行のプロファイルを添付します。

プロファイル、最初のインスタンス

プロファイル、2 番目のインスタンス

ここで、同じシナリオの動作を示したいと思いますが、matrixMul サンプルの 2 つのインスタンスを同時に実行します。

プロファイル、最初のインスタンス

プロファイル、2 番目のインスタンス

ご覧のとおり、最初のシナリオでは、カーネルが他のカーネルの終了を待ちます。一方、2 番目のシナリオ (matrixMul) では、両方のコンテキストのカーネルが同時に実行されます。

ありがとうございました。