6
Ich programmiere OpenCL mit einer GeForce GT 610 Karte in Linux. Meine CPU und GPU Ergebnisse mit doppelter Genauigkeit sind nicht konsistent. Ich kann hier einen Teil des Codes posten, aber ich möchte zuerst wissen, ob jemand anderes mit diesem Problem konfrontiert ist. Der Unterschied zwischen den Ergebnissen der GPU- und CPU-Doppelpräzision wird deutlich, wenn ich Schleifen mit vielen Iterationen durchführe. Es gibt wirklich nichts besonderes an dem Code, aber ich kann es hier posten, wenn jemand interessiert ist. Danke vielmals. Hier ist mein Code. Bitte entschuldigen Sie die __ und schlechte Formatierung, da ich neu hier bin :) Wie Sie sehen können, habe ich zwei Schleifen und mein CPU-Code ist im Wesentlichen fast eine identische Version.OpenCL double precision unterscheidet sich von CPU double precision
#ifdef cl_khr_fp64
#pragma OPENCL EXTENSION cl_khr_fp64 : enable
#elif defined(cl_amd_fp64)
#pragma OPENCL EXTENSION cl_amd_fp64 : enable
#else
#error "Double precision floating point not supported by OpenCL implementation."
#endif
__kernel void simpar(__global double* fp, __global double* fp1,
__global double* fp3, __global double* fp5,
__global double* fp6, __global double* fp7,
__global double* fp8, __global double* fp8Plus,
__global double* x, __global double* v, __global double* acc,
__global double* keBuf, __global double* peBuf,
unsigned int prntstps, unsigned int nprntstps, double dt
) {
unsigned int m,i,j,k,l,t;
unsigned int chainlngth=100;
double dxi, twodxi, dxipl1, dximn1, fac, fac1, fac2, fac13, fac23;
double ke,pe,tke,tpe,te,dx;
double hdt, hdt2;
double alpha=0.16;
double beta=0.7;
double cmass;
double peTemp;
nprntstps=1001;
dt=0.01;
prntstps=100;
double alphaby4=beta/4.0;
hdt=0.5*dt;
hdt2=dt*0.5*dt;
double Xlocal,Vlocal,Acclocal;
unsigned int global_id=get_global_id(0);
if (global_id<chainlngth){
Xlocal=x[global_id];
Vlocal=v[global_id];
Acclocal=acc[global_id];
for (m=0;m<nprntstps;m++){
for(l=0;l<prntstps;l++){
Xlocal =Xlocal+dt *Vlocal+hdt2*Acclocal;
x[global_id]=Xlocal;
barrier(CLK_LOCAL_MEM_FENCE);
Vlocal =Vlocal+ hdt * Acclocal;
barrier(CLK_LOCAL_MEM_FENCE);
j = global_id - 1;
k = global_id + 1;
if (j == -1) {
dximn1 = 0.0;
} else {
dximn1 = x[j];
}
if (k == chainlngth) {
dxipl1 = 0.0;
} else {
dxipl1 = x[k];
}
dxi = Xlocal;
twodxi = 2.0 * dxi;
fac = dxipl1 + dximn1 - twodxi;
fac1 = dxipl1 - dxi;
fac2 = dxi - dximn1;
fac13 = fac1 * fac1 * fac1;
fac23 = fac2 * fac2 * fac2;
Acclocal = alpha * fac + beta * (fac13 - fac23);
barrier(CLK_GLOBAL_MEM_FENCE);
Vlocal += hdt * Acclocal;
v[global_id]=Vlocal;
acc[global_id]=Acclocal;
barrier(CLK_GLOBAL_MEM_FENCE);
}
barrier(CLK_GLOBAL_MEM_FENCE);
tke = tpe = te = dx = 0.0;
ke=0.5*Vlocal*Vlocal;//Vlocal*Vlocal;
barrier(CLK_GLOBAL_MEM_FENCE);
fp6[(m*100)+global_id]=ke;
keBuf[global_id]=ke;
ke=0.0;
barrier(CLK_GLOBAL_MEM_FENCE);
j = global_id - 1;
k = global_id + 1;
if (j == -1) {
dximn1 = 0.0;
} else {
dximn1 = x[j];
}
if (k == chainlngth) {
dxipl1 = 0.0;
} else {
dxipl1 = x[k];
}
dxi = Xlocal;
twodxi = 2.0 * dxi;
fac = dxipl1 + dximn1 - twodxi;
fac1 = dxipl1 - dxi;
fac2 = dxi - dximn1;
fac13 = fac1 * fac1 * fac1;
fac23 = fac2 * fac2 * fac2;
Acclocal = alpha * fac + beta * (fac13 - fac23);
barrier(CLK_GLOBAL_MEM_FENCE);
Vlocal += hdt * Acclocal;
v[global_id]=Vlocal;
acc[global_id]=Acclocal;
barrier(CLK_GLOBAL_MEM_FENCE);
}
barrier(CLK_GLOBAL_MEM_FENCE);
tke = tpe = te = dx = 0.0;
ke=0.5*Vlocal*Vlocal;//Vlocal*Vlocal;
barrier(CLK_GLOBAL_MEM_FENCE);
fp6[(m*100)+global_id]=ke;
keBuf[global_id]=ke;
ke=0.0;
barrier(CLK_GLOBAL_MEM_FENCE);
j = global_id - 1;
k = global_id + 1;
if (j == -1) {
dximn1 = 0.0;
} else {
dximn1 = x[j];
}
if (k == chainlngth) {
dxipl1 = 0.0;
} else {
dxipl1 = x[k];
}
dxi = Xlocal;
twodxi = 2.0 * dxi;
fac = dxipl1 + dximn1 - twodxi;
fac1 = dxipl1 - dxi;
fac2 = dxi - dximn1;
fac13 = fac1 * fac1 * fac1;
fac23 = fac2 * fac2 * fac2;
Acclocal = alpha * fac + beta * (fac13 - fac23);
barrier(CLK_GLOBAL_MEM_FENCE);
Vlocal += hdt * Acclocal;
v[global_id]=Vlocal;
acc[global_id]=Acclocal;
barrier(CLK_GLOBAL_MEM_FENCE);
}
barrier(CLK_GLOBAL_MEM_FENCE);
tke = tpe = te = dx = 0.0;
ke=0.5*Vlocal*Vlocal;//Vlocal*Vlocal;
barrier(CLK_GLOBAL_MEM_FENCE);
fp6[(m*100)+global_id]=ke;
keBuf[global_id]=ke;
ke=0.0;
barrier(CLK_GLOBAL_MEM_FENCE);
if (global_id ==0){
for(t=0;t<100;t++)
tke+=keBuf[t];
}
barrier(CLK_GLOBAL_MEM_FENCE);
k = global_id-1;
if (k == -1) {
dx = Xlocal;
}else{
dx = Xlocal-x[k];
}
fac = dx * dx;
peTemp = alpha * 0.5 * fac + alphaby4 * fac * fac;
fp8[global_id*m]=peTemp;
if (global_id == 0)
tpe+=peTemp;
barrier(CLK_GLOBAL_MEM_FENCE);
cmass=0.0;
dx = -x[100-1];
fac = dx*dx;
pe=alpha*0.5*fac+alphaby4*fac*fac;
if (global_id==0){
fp8Plus[m]=pe;
tpe+=peBuf[0];
fp5[m*2]=i;
fp5[m*2+1]=cmass;
te=tke+tpe;
fp[m*2]=m;
fp[m*2+1]=te;
}
barrier(CLK_GLOBAL_MEM_FENCE);
//cmass /=100;
fp1[(m*chainlngth)+global_id]=Xlocal-cmass;
// barrier(CLK_GLOBAL_MEM_FENCE);
fp3[(m*chainlngth)+global_id]=Vlocal;
// barrier(CLK_GLOBAL_MEM_FENCE);
fp7[(m*chainlngth)+global_id]=Acclocal;
barrier(CLK_GLOBAL_MEM_FENCE);
}
}
}
Wenn Sie "nicht konsistent" sagen, was meinen Sie eigentlich? Können Sie die Unterschiede quantitativ darstellen (absolute und relative Fehler, Anzahl der letzten Stellen in Übereinstimmung)? – talonmies
Ich führe die gleichen (zwei) verschachtelte Schleifen. Die innere Schleife wird 100 Mal ausgeführt, und dann druckt die äußere Schleife einige Ergebnisse. Hier wird eine Probe aus der ersten Iteration der äußeren Schleife: CPU: 0,0000000011 0,0000030832 0,0005244239 0,1400572807 1,5213598941 1,5213598941 0,1400572807 0,0005244239 0,0000030832 0,0000000011 GPU: 0,0000000012 0,0000032002 0,0005183133 0,1401775827 1,5249561626 1,5249561626 0,1401775827 0,0005183133 0.0000032002 0.0000000012 – Newbee
Wenn Sie "CPU double precision" sagen, meinen Sie, dass Sie den gleichen Kernel auf dem CPU-Gerät ausführen, oder vergleichen Sie ihn mit einer nativen Implementierung? Stellen Sie außerdem sicher, dass Sie keine Flags an den OpenCL-Compiler übergeben, was die IEEE-Strenge (z. B. das Aktivieren von schnellen Mehrfachadditionen usw.) erleichtern könnte. Und ja, der Code würde sicherlich dazu beitragen, dass Leute sich reproduzieren und ihre eigenen Tests durchführen können, ohne Code ist das Beste, was wir tun können, Spekulationen. – Thomas