OpenCL nested l oop дает неожиданные результаты - PullRequest
Новая
       130

OpenCL nested l oop дает неожиданные результаты

0 голосов
/ 21 января 2020

Я немного новичок в OpenCL / C99, но не могу понять, почему два приведенных ниже ядра дают разные результаты. X был инициализирован нулями, но требует «обнуления» на каждом внешнем шаге l oop, иначе получаются неправильные результаты (см. График). Обратите внимание, что я не использовал здесь никакого параллелизма; насколько мне известно, это полностью серийно:

содержание kernels.cl: 

#pragma OPENCL EXTENSION cl_khr_fp64: enable
#define __CL_ENABLE_EXCEPTIONS

__kernel void nested_sum_succeeds(
    int L,
    __global __read_only double* X,
    __global double* Y
)
{
    for (int k=0; k<=L; k++) {
        Y[k] = 0;
        for (int j=0; j<=k; j++) {
            Y[k] += X[j];
        }
    }
}

__kernel void nested_sum_fails(
    int L,
    __global __read_only double* X,
    __global double* Y
)
{
    for (int k=0; k<=L; k++) {
        // Y[k] = 0;
        for (int j=0; j<=k; j++) {
            Y[k] += X[j];
        }
    }
}

содержание script.py: 

import numpy as np
import pyopencl as cl
import pyopencl.array as cl_array
import matplotlib.pyplot as plt

with open("./kernels.cl") as fp:
    prog_str = fp.read()
ctx = cl.create_some_context()
queue = cl.CommandQueue(ctx)
prog = cl.Program(ctx, prog_str).build()

L = 1000
X = np.linspace(0, 10, L)
X_dev = cl_array.to_device(queue, X)
Y_succeeds_dev = cl_array.to_device(queue, np.zeros(shape=X.shape, dtype=np.float64))
Y_fails_dev = cl_array.to_device(queue, np.zeros(shape=X.shape, dtype=np.float64))


nested_sum_succeeds = prog.nested_sum_succeeds
nested_sum_succeeds.set_scalar_arg_dtypes([
    np.int64,
    None,
    None,
])

nested_sum_succeeds(
    queue,
    (len(X),),
    None,
    L,
    X_dev.data,
    Y_succeeds_dev.data,
)


nested_sum_fails = prog.nested_sum_fails
nested_sum_fails.set_scalar_arg_dtypes([
    np.int64,
    None,
    None,
])

nested_sum_fails(
    queue,
    (len(X),),
    None,
    L,
    X_dev.data,
    Y_fails_dev.data,
)

np.allclose(Y_succeeds_dev.get(), Y_fails_dev.get()) #False

plt.ion()
plt.plot(Y_succeeds_dev.get())
plt.plot(Y_fails_dev.get())

Результаты:

enter image description here

1 Ответ

3 голосов
/ 21 января 2020

Обратите внимание, что я не использовал здесь никакого параллелизма;

Да, ядро ​​работает параллельно - планируется запустить len(x) рабочих элементов в первом измерении. Как только вы измените его на обработку, используя 1 рабочий элемент, все в порядке: 

nested_sum_succeeds(
    queue,
    (1,),
    None,
    L,
    X_dev.data,
    Y_succeeds_dev.data,
)

nested_sum_fails(
    queue,
    (1,),
    None,
    L,
    X_dev.data,
    Y_fails_dev.data,
)

Затем np.allclose(Y_succeeds_dev.get(), Y_fails_dev.get()) вернет True. Вы также можете удалить обнуление Y[k] = 0; из ядра nested_sum_succeeds, так как оно не нужно.

Также, если вы хотите запустить это ядро ​​на других устройствах, вам понадобятся некоторые незначительные исправления, потому что не все компиляторы собираются принять, что тип первого аргумента ядра находится в ядре int и в python указан как np.int64, что должно соответствовать тому, что находится в ядре, так: 

nested_sum_succeeds.set_scalar_arg_dtypes([
    np.int32,
    None,
    None,
])

nested_sum_fails.set_scalar_arg_dtypes([
    np.int32,
    None,
    None,
])

И Еще одна вещь, которая относится к использованию на других устройствах, я бы удалил __read_only квалификатор доступа, который также не будет компилироваться на всех устройствах.

...