0

CUDA を使用して次の補間を開発しました。この補間を改善する方法を探しています。何らかの理由で、CUDA テクスチャを使用したくありません。

いくつかの未知の理由で気づいたもう 1 つの点は、ベクトルのサイズがスレッドの数よりも優れている場合 (たとえば、サイズ 1000 のベクトルの場合、私の場合はベクトル全体で補間が実行されないことです。 512 に等しいスレッド数. スレッドは最初のジョブを実行し、それだけです. singleInterp 関数を最適化したいと思います.

これが私のコードです：

__device__ float singleInterp(float* data, float x, int lx_data) { 

float res = 0;    
int i1=0;
int j=lx_data; 
int imid;

while (j>i1+1)  
{  
    imid = (int)(i1+j+1)/2;
    if (data[imid]<x) 
         i1=imid;
    else 
        j=imid;
} 
if (i1==j)
    res = data[i1+lx_data];
else
    res =__fmaf_rn( __fdividef(data[j+lx_data]-data[i1+lx_data],(data[j]-data[i1])),x-data[i1], data[i1+lx_data]);

return res;

}

カーネル：

__global__ void linearInterpolation(float* data, float* x_in, int lx_data) {

int i = threadIdx.x + blockDim.x * blockIdx.x; 
int index = i; 
if (index < lx_data) 
    x_in[index] = singleInterp(data, x_in[index], lx_data);
}
4

1 に答える 1

1

1D 線形補間に興味があるようです。私はすでにこの種の補間を最適化するという問題を抱えていましたが、次のコードになりました

__global__ void linear_interpolation_kernel_function_GPU(double* __restrict__ result_d, const double* __restrict__ data_d, const double* __restrict__ x_out_d, const int M, const int N)
{
    int j = threadIdx.x + blockDim.x * blockIdx.x;

    if(j<N)
    {
        double reg_x_out = x_out_d[j/2]+M/2;
        int k = floor(reg_x_out);
        double a = (reg_x_out)-floor(reg_x_out);
        double dk = data_d[2*k+(j&1)];
        double dkp1 = data_d[2*k+2+(j&1)];
        result_d[j] = a * dkp1 + (-dk * a + dk);
    } 
 }

-M/2データは、との間の整数ノードでサンプリングされると想定されますM/2。次のWeb ページで説明されているように、このコードは 1D テクスチャ補間と「同等」です。1D 線形テクスチャ補間については、CUDA-Programming-Guideの Fig. 13 を参照してください。異なるソリューション間の比較については、次のスレッドを参照してください。

于 2013-03-07T16:57:14.207 に答える