OpenACC представляет данные об обновлении предложения

Question

OpenACC представляет данные об обновлении предложения

Я пытаюсь сделать оптимизацию openACC для многих симуляций тела. В настоящее время я сталкиваюсь с проблемой, которая приводит к проблеме с памятью ниже

позвонить cuStreamSynchronize вернула ошибку 700: Неверный адрес во время выполнения ядра
позвонить cuMemFreeHost вернула ошибку 700: Неверный адрес во время выполнения ядра
srun: ошибка: jrc0017: задача 0: завершена с кодом выхода 1

я использую pgc++ компилятор и мои флаги компилятора -acc -Minfo=accel -ta=tesla -fast -std=c++11 и я не хочу использовать -ta=tesla:managed потому что я хочу организовать память самостоятельно.

#pragma acc kernels present(sim.part.rx, sim.part.ry, sim.part.rz, sim.part.vx, sim.part.vy, sim.part.vz)
{
for(int idx = 0; idx < sim.num; ++idx) {     // Loop over target particle
float
prx = sim.part.rx[idx],                   // my position
pry = sim.part.ry[idx],
prz = sim.part.rz[idx];
float Fx = 0.f, Fy = 0.f, Fz = 0.f;          // Force
#pragma acc loop
for(int jdx = 0; jdx < sim.num; ++jdx) {   // Loop over interaction partners
if(idx != jdx) {                          // No self-force
const float dx = prx - sim.part.rx[jdx]; // Distance to partner
const float dy = pry - sim.part.ry[jdx];
const float dz = prz - sim.part.rz[jdx];
const float h  = 1.f/sqrt(dx*dx + dy*dy + dz*dz + eps);
const float h3 = h*h*h;
Fx += dx*h3;                            // Sum up force
Fy += dy*h3;
Fz += dz*h3;
}
}
sim.part.vx[idx] += sim.mass*dt*Fx;         // update velocity
sim.part.vy[idx] += sim.mass*dt*Fy;
sim.part.vz[idx] += sim.mass*dt*Fz;
}
}

Если я удалю код ниже

sim.part.vx[idx] += sim.mass*dt*Fx;         // update velocity
sim.part.vy[idx] += sim.mass*dt*Fy;
sim.part.vz[idx] += sim.mass*dt*Fz;

мой код может работать без проблем. Но у меня проблемы с памятью, если я откомментирую их. Кажется, что sim.part.vx пытаются обновить данные, но компилятор не знает, что приводит к проблеме с памятью.

Кто-нибудь знает, как решить эту проблему?

1

c++c++11 gpu openacc

Решение

Другие решения

Других решений пока нет …

Источник

Accepted Answer

Я подозреваю, что проблема в том, что sim а также sim.part не на устройстве (или компилятор не понимает, что они на устройстве. В качестве обходного пути, вы можете попробовать ввести указатели на эти массивы напрямую?

float *rx = sim.part.rx, *ry = sim.part.ry, *rz = sim.part.rz,
*vx = sim.part.vx, *vy = sim.part.vy, *vz = sim.part.vz;
#pragma acc kernels present(rx, ry, rz, vx, vy, vz)
{
for(int idx = 0; idx < sim.num; ++idx) {     // Loop over target particle
float
prx = rx[idx],                   // my position
pry = ry[idx],
prz = rz[idx];
float Fx = 0.f, Fy = 0.f, Fz = 0.f;          // Force
#pragma acc loop
for(int jdx = 0; jdx < sim.num; ++jdx) {   // Loop over interaction partners
if(idx != jdx) {                          // No self-force
const float dx = prx - rx[jdx]; // Distance to partner
const float dy = pry - ry[jdx];
const float dz = prz - rz[jdx];
const float h  = 1.f/sqrt(dx*dx + dy*dy + dz*dz + eps);
const float h3 = h*h*h;
Fx += dx*h3;                            // Sum up force
Fy += dy*h3;
Fz += dz*h3;
}
}
vx[idx] += sim.mass*dt*Fx;         // update velocity
vy[idx] += sim.mass*dt*Fy;
vz[idx] += sim.mass*dt*Fz;
}
}

Как распределяются sim и sim.part? Можно использовать директивы неструктурированных данных в конструкторе и деструкторе, чтобы убедиться, что sim и sim.part также находятся на устройстве. Если вы уже сделали это, то еще одно возможное решение — добавить present(sim, sim.part) к вашему существующему предложению, чтобы компилятор знал, что вы уже позаботились и об этих структурах данных.

2