cuda - CUDAの共有メモリを解放します

Question

CUDAのCTAによって割り当てられた共有メモリを解放するために利用できるアプリケーションレベルのAPIはありますか？CTAを別のタスクに再利用したいので、そのタスクを開始する前に、前のタスクで使用されていたメモリをクリーンアップする必要があります。

Answer 1

共有メモリは、カーネルの起動時に静的に割り当てられます。オプションで、カーネルでサイズなしの共有割り当てを指定できます。

__global__ void MyKernel()
{
    __shared__ int fixedShared;
    extern __shared__ int extraShared[];
    ...
}

次に、3番目のカーネル起動パラメーターは、そのサイズ設定されていない割り当てに対応する共有メモリの量を指定します。

MyKernel<<<blocks, threads, numInts*sizeof(int)>>>( ... );

カーネルの起動に割り当てられる共有メモリの合計量は、カーネルで宣言された量、共有メモリカーネルパラメータ、およびアライメントオーバーヘッドの合計です。それを「解放」することはできません。カーネルの起動中は割り当てられたままになります。

実行の複数のフェーズを経て、さまざまな目的で共有メモリを使用する必要があるカーネルの場合、実行できることは、共有メモリポインタでメモリを再利用することです。サイズなしの宣言でポインタ演算を使用します。

何かのようなもの：

__global__ void MyKernel()
{
    __shared__ int fixedShared;
    extern __shared__ int extraShared[];
    ...
    __syncthreads();
    char *nowINeedChars = (char *) extraShared;
    ...
}

このイディオムを使用するSDKサンプルはわかりませんが、threadFenceReductionサンプルはを宣言し、__shared__ bool共有メモリを使用して削減の部分的な合計を保持します。