cuda - cudaで3Dマトリックスを効率的にシフトする方法

Question

GPU に 3D マトリックス M[i*strideyz+j*stridez+k] があり、このマトリックス レイヤーをレイヤーごとにシフトしたい: i+1-->i で、新しいデータを最後のレイヤーに追加します。そのためにカーネルを書きましたが、非常に遅いです。ここでいくつかの議論があることは知っていますが、どういうわけか答えが見つかりません。

これが私の実装です：

__global void shift(int nlayers, float* M, float* inp,
                    size_t strideyz,size_t stridez)
{
  int k=blockIdx.x*blockDim.x+threadIdx.x;
  int j=blockIdx.y*blockDim.y+threadIdx.y;
  for(i=0;i<nlayers-1;i++)
    M[i*strideyz+j*stridez+k]=M[(i+1)*strideyz+j*stridez+k];
  M[(Nlayers-1)*strideyz+j*stridez+k]=inp[j*stridez+k];
}