cuda - カーネル最適化

Question

私は単純なパーティクルシステムを作成しました。私は計算能力2.1のデバイスを持っています。カーネルを最適化するために私は何を変更できますか？

変数tPosとtVelがレジスタに格納されているとします。

__global__ void particles_kernel(float4 *vbo, float4 *pos, float4 *vel) 
{ 
    int tid = blockIdx.x * blockDim.x + threadIdx.x; 

    float4 tPos = pos[tid]; 
    float4 tVel = vel[tid]; 

    tPos.x += tVel.x; 
    tPos.y += tVel.y; 
    tPos.z += tVel.z; 

    if(tPos.x < -2.0f) 
    { 
     tVel.x = -tVel.x; 
    } 
    else if(tPos.x > 2.0f) 
    { 
     tVel.x = -tVel.x; 
    } 


    if(tPos.y < -2.0f) 
    { 
     tVel.y = -tVel.y; 
    } 
    else if(tPos.y > 2.0f) 
    { 
     tVel.y = -tVel.y; 
    } 


    if(tPos.z < -2.0f) 
    { 
     tVel.z = -tVel.z; 
    } 
    else if(tPos.z > 2.0f) 
    { 
     tVel.z = -tVel.z; 
    } 


    pos[tid] = tPos; 
    vel[tid] = tVel; 


    vbo[tid] = make_float4(tPos.x, tPos.y, tPos.z, tPos.w); 
} 

Answer 1

私は何かが欠けていない限り、あなたの締め付けコードは次のように簡略化することができます。コードはに表示される

if (fabsf(tVel.x) > 2.0f) tVel.x = -tVel.x; 
if (fabsf(tVel.y) > 2.0f) tVel.y = -tVel.y; 
if (fabsf(tVel.z) > 2.0f) tVel.z = -tVel.z; 

しかし、計算の比較的小さなアモン与えられ、この変更はおそらく、パフォーマンスが改善されませんメモリに束縛されている（データをストリーミングしている）。たぶんあなたのアプリに計算の密度を上げるためにこの計算と組み合わせることができる別の計算がありますか？