OpenCL カーネルによる内部レジスタの使用に関して、少し混乱しています。-cl-nv-verbose を使用して、カーネルのレジスタの使用状況をキャプチャしています。現在、私のカーネルはカーネルptxas info: Used 4 registers
内のいくつかのコードを記録しています。次のセグメントの場合:
double a;
a = pow(2.0,2.0)
if (index != 0) {
}
使用するレジスタが変わりますptxas info: Used 6 registers
。if ループには何もないことがわかりました。しかし、次のように再構成すると:
double a;
if (index != 0) {
a = pow(2.0,2.0)
}
これにより、レジスタの使用法が に変更されますptxas info: Used 15 registers
。カーネルのワーク グループ サイズは変更しません。おそらく答えはptxコードを見ることにあるのですが、私はそれを理解していません(必要に応じて取得できますが)。私がもっと興味を持っているのは、コード行を移動するだけで、レジスタの使用量が 2 倍になる理由です。何か案は?(インデックスはprivate
)
更新: カーネルコード:
__kernel void butterfC( __global double *sI,
__global double *sJ,
__global double *sK,
const int zR,
const int yR,
const int xR,
unsigned int l1,
const int dir,
unsigned int type )
{
int idX = get_global_id(0);
int idY = get_global_id(1);
int idZ = get_global_id(2);
int BASE = idZ*xR*yR;
int STRIDE = 1;
int powX = pow(4.0f,l1);
int powXm1 = pow(4.0f,l1-1);
int yIndex, kIndex;
switch(type)
{
case 1: BASE += idY*xR;
yIndex = idX / powXm1 * powX;
kIndex = (idX % powXm1) + yIndex;
break;
case 2: BASE += idX; STRIDE = xR;
yIndex = idY / powXm1 * powX;
kIndex = idY % powXm1 + yIndex;
break;
case 3: BASE = idY*xR + idX; STRIDE = xR * yR;
yIndex = idZ / powXm1 * powX;
kIndex = idZ % powXm1 + yIndex;
break;
}
double a;
//a = pow(2.0,2.0);
if (kIndex != 0) {
a = pow(2.0,2.0);
.... do stuff
}
}