OpenCL & Java - странные результаты производительности
Я пытаюсь изучить использование OpenCL для повышения производительности некоторого кода Java с использованием JOCL. Я просматривал примеры, представленные на их веб-сайте, и использовал их, чтобы собрать быструю программу, чтобы сравнить ее производительность с работой в обычном режиме. Результаты, которые я получаю, немного неожиданны, и я обеспокоен, что могу делать что-то не так.
Для начала я использую JOCL 0.1.9, так как у меня есть карта NVIDIA, которая не поддерживает OpenCL/JOCL 2.0. Мой компьютер имеет процессор Intel Core i7, карту Intel HD Graphics 530 и NVIDIA Quadro M2000M.
Программа, которую я написал, основана на примерах JOCL; он берет два массива чисел и умножает их, помещая результаты в третий массив. Я использую метод Java nanoTime(), чтобы примерно отслеживать наблюдаемое время выполнения Java.
public class PerformanceComparison {
public static final int ARRAY_SIZE = 1000000;
// OpenCL kernel code
private static String programSource = "__kernel void " + "sampleKernel(__global const float *a,"
+ " __global const float *b," + " __global float *c)" + "{"
+ " int gid = get_global_id(0);" + " c[gid] = a[gid] * b[gid];" + "}";
public static final void main(String[] args) {
// build arrays
float[] sourceA = new float[ARRAY_SIZE];
float[] sourceB = new float[ARRAY_SIZE];
float[] nvidiaResult = new float[ARRAY_SIZE];
float[] intelCPUResult = new float[ARRAY_SIZE];
float[] intelGPUResult = new float[ARRAY_SIZE];
float[] javaResult = new float[ARRAY_SIZE];
for (int i = 0; i < ARRAY_SIZE; i++) {
sourceA[i] = i;
sourceB[i] = i;
}
// get platforms
cl_platform_id[] platforms = new cl_platform_id[2];
clGetPlatformIDs(2, platforms, null);
// I know what devices I have, so declare variables for each of them
cl_context intelCPUContext = null;
cl_context intelGPUContext = null;
cl_context nvidiaContext = null;
cl_device_id intelCPUDevice = null;
cl_device_id intelGPUDevice = null;
cl_device_id nvidiaDevice = null;
// get all devices on all platforms
for (int i = 0; i < 2; i++) {
cl_platform_id platform = platforms[i];
cl_context_properties properties = new cl_context_properties();
properties.addProperty(CL_CONTEXT_PLATFORM, platform);
int[] numDevices = new int[1];
cl_device_id[] devices = new cl_device_id[2];
clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 2, devices, numDevices);
// get devices and build contexts
for (int j = 0; j < numDevices[0]; j++) {
cl_device_id device = devices[j];
cl_context context = clCreateContext(properties, 1, new cl_device_id[] { device }, null, null, null);
long[] length = new long[1];
byte[] buffer = new byte[2000];
clGetDeviceInfo(device, CL_DEVICE_NAME, 2000, Pointer.to(buffer), length);
String deviceName = new String(buffer, 0, (int) length[0] - 1);
// save based on the device name
if (deviceName.contains("Quadro")) {
nvidiaContext = context;
nvidiaDevice = device;
}
if (deviceName.contains("Core(TM)")) {
intelCPUContext = context;
intelGPUDevice = device;
}
if (deviceName.contains("HD Graphics")) {
intelGPUContext = context;
intelGPUDevice = device;
}
}
}
// multiply the arrays using Java and on each of the devices
long jvmElapsed = runInJVM(sourceA, sourceB, javaResult);
long intelCPUElapsed = runInJOCL(intelCPUContext, intelCPUDevice, sourceA, sourceB, intelCPUResult);
long intelGPUElapsed = runInJOCL(intelGPUContext, intelGPUDevice, sourceA, sourceB, intelGPUResult);
long nvidiaElapsed = runInJOCL(nvidiaContext, nvidiaDevice, sourceA, sourceB, nvidiaResult);
// results
System.out.println("Standard Java Runtime: " + jvmElapsed + " ns");
System.out.println("Intel CPU Runtime: " + intelCPUElapsed + " ns");
System.out.println("Intel GPU Runtime: " + intelGPUElapsed + " ns");
System.out.println("NVIDIA GPU Runtime: " + nvidiaElapsed + " ns");
}
/**
* The basic Java approach - loop through the arrays, and save their results into the third array
*
* @param sourceA multiplicand
* @param sourceB multiplier
* @param result product
* @return the (rough) execution time in nanoseconds
*/
private static long runInJVM(float[] sourceA, float[] sourceB, float[] result) {
long startTime = System.nanoTime();
for (int i = 0; i < ARRAY_SIZE; i++) {
result[i] = sourceA[i] * sourceB[i];
}
long endTime = System.nanoTime();
return endTime - startTime;
}
/**
* Run a more-or-less equivalent program in OpenCL on the specified device
*
* @param context JOCL context
* @param device JOCL device
* @param sourceA multiplicand
* @param sourceB multiplier
* @param result product
* @return the (rough) execution time in nanoseconds
*/
private static long runInJOCL(cl_context context, cl_device_id device, float[] sourceA, float[] sourceB,
float[] result) {
// create command queue
cl_command_queue commandQueue = clCreateCommandQueue(context, device, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, null);
// allocate memory
cl_mem memObjects[] = new cl_mem[3];
memObjects[0] = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, Sizeof.cl_float * ARRAY_SIZE,
Pointer.to(sourceA), null);
memObjects[1] = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, Sizeof.cl_float * ARRAY_SIZE,
Pointer.to(sourceB), null);
memObjects[2] = clCreateBuffer(context, CL_MEM_READ_WRITE, Sizeof.cl_float * ARRAY_SIZE, null, null);
// build program and set arguments
cl_program program = clCreateProgramWithSource(context, 1, new String[] { programSource }, null, null);
clBuildProgram(program, 0, null, null, null, null);
cl_kernel kernel = clCreateKernel(program, "sampleKernel", null);
clSetKernelArg(kernel, 0, Sizeof.cl_mem, Pointer.to(memObjects[0]));
clSetKernelArg(kernel, 1, Sizeof.cl_mem, Pointer.to(memObjects[1]));
clSetKernelArg(kernel, 2, Sizeof.cl_mem, Pointer.to(memObjects[2]));
long global_work_size[] = new long[]{ARRAY_SIZE};
long local_work_size[] = new long[]{1};
// Execute the kernel
long startTime = System.nanoTime();
clEnqueueNDRangeKernel(commandQueue, kernel, 1, null,
global_work_size, local_work_size, 0, null, null);
// Read the output data
clEnqueueReadBuffer(commandQueue, memObjects[2], CL_TRUE, 0,
ARRAY_SIZE * Sizeof.cl_float, Pointer.to(result), 0, null, null);
long endTime = System.nanoTime();
// Release kernel, program, and memory objects
clReleaseMemObject(memObjects[0]);
clReleaseMemObject(memObjects[1]);
clReleaseMemObject(memObjects[2]);
clReleaseKernel(kernel);
clReleaseProgram(program);
clReleaseCommandQueue(commandQueue);
clReleaseContext(context);
return endTime - startTime;
}
}
Выход программы:
Standard Java Runtime: 3662913 ns
Intel CPU Runtime: 27186 ns
Intel GPU Runtime: 9817 ns
NVIDIA GPU Runtime: 12400512 ns
Об этом меня смущают две вещи:
- Почему программа работает намного быстрее на процессоре при использовании OpenCL? Это то же оборудование, которое будет использовать JVM; Я знаю, что Java медленная по сравнению с языками более низкого уровня, такими как OpenCL, но я не думала, что она такая медленная.
- Что не так с картой NVIDIA? Я знаю, что их поддержка OpenCL менее чем звездна, учитывая их среду CUDA, но я все же ожидал бы, что она будет по крайней мере быстрее, чем обычно. На самом деле, резервная копия, "это здесь, в случае, если вы сломаете свою настоящую графическую карту", Intel GPU вращается вокруг нее.
Я беспокоюсь, что я делаю что-то не так или, по крайней мере, упускаю что-то, что позволит этому работать в полную силу. Любые указатели, которые я мог получить, были бы очень кстати.
PS - я знаю, что, поскольку у меня есть карта NVIDIA, CUDA, вероятно, будет лучшим / быстрым вариантом для меня; однако в этом случае я бы предпочел гибкость OpenCL.
Обновление: я смог найти одну вещь, которую я сделал неправильно; полагаться на Java, чтобы сообщить, что время выполнения было тупым. Я написал новый тест с использованием профилирования OpenCL, и он дает немного более ощутимые результаты:
Код:
public class PerformanceComparisonTakeTwo {
//@formatter:off
private static final String PROFILE_TEST =
"__kernel void "
+ "sampleKernel(__global const float *a,"
+ " __global const float *b,"
+ " __global float *c,"
+ " __global float *d,"
+ " __global float *e,"
+ " __global float *f)"
+ "{"
+ " int gid = get_global_id(0);"
+ " c[gid] = a[gid] + b[gid];"
+ " d[gid] = a[gid] - b[gid];"
+ " e[gid] = a[gid] * b[gid];"
+ " f[gid] = a[gid] / b[gid];"
+ "}";
//@formatter:on
private static final int ARRAY_SIZE = 100000000;
public static final void main(String[] args) {
initialize();
}
public static void initialize() {
// identify all platforms
cl_platform_id[] platforms = getPlatforms();
Map<cl_device_id, cl_platform_id> deviceMap = getDevices(platforms);
performProfilingTest(deviceMap);
}
private static cl_platform_id[] getPlatforms() {
int[] platformCount = new int[1];
clGetPlatformIDs(0, null, platformCount);
cl_platform_id[] platforms = new cl_platform_id[platformCount[0]];
clGetPlatformIDs(platforms.length, platforms, platformCount);
return platforms;
}
private static Map<cl_device_id, cl_platform_id> getDevices(cl_platform_id[] platforms) {
Map<cl_device_id, cl_platform_id> deviceMap = new HashMap<>();
for(int i = 0; i < platforms.length; i++) {
int[] deviceCount = new int[1];
clGetDeviceIDs(platforms[i], CL_DEVICE_TYPE_ALL, 0, null, deviceCount);
cl_device_id[] devices = new cl_device_id[deviceCount[0]];
clGetDeviceIDs(platforms[i], CL_DEVICE_TYPE_ALL, devices.length, devices, null);
for(int j = 0; j < devices.length; j++) {
deviceMap.put(devices[j], platforms[i]);
}
}
return deviceMap;
}
private static void performProfilingTest(Map<cl_device_id, cl_platform_id> deviceMap) {
float[] sourceA = new float[ARRAY_SIZE];
float[] sourceB = new float[ARRAY_SIZE];
for(int i = 0; i < ARRAY_SIZE; i++) {
sourceA[i] = i;
sourceB[i] = i;
}
for(Entry<cl_device_id, cl_platform_id> devicePair : deviceMap.entrySet()) {
cl_device_id device = devicePair.getKey();
cl_platform_id platform = devicePair.getValue();
cl_context_properties properties = new cl_context_properties();
properties.addProperty(CL_CONTEXT_PLATFORM, platform);
cl_context context = clCreateContext(properties, 1, new cl_device_id[] { device }, null, null, null);
cl_command_queue commandQueue = clCreateCommandQueue(context, device, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE | CL_QUEUE_PROFILING_ENABLE, null);
cl_mem memObjects[] = new cl_mem[6];
memObjects[0] = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, Sizeof.cl_float * ARRAY_SIZE,
Pointer.to(sourceA), null);
memObjects[1] = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, Sizeof.cl_float * ARRAY_SIZE,
Pointer.to(sourceB), null);
memObjects[2] = clCreateBuffer(context, CL_MEM_READ_WRITE, Sizeof.cl_float * ARRAY_SIZE, null, null);
memObjects[3] = clCreateBuffer(context, CL_MEM_READ_WRITE, Sizeof.cl_float * ARRAY_SIZE, null, null);
memObjects[4] = clCreateBuffer(context, CL_MEM_READ_WRITE, Sizeof.cl_float * ARRAY_SIZE, null, null);
memObjects[5] = clCreateBuffer(context, CL_MEM_READ_WRITE, Sizeof.cl_float * ARRAY_SIZE, null, null);
cl_program program = clCreateProgramWithSource(context, 1, new String[] { PROFILE_TEST }, null, null);
clBuildProgram(program, 0, null, null, null, null);
cl_kernel kernel = clCreateKernel(program, "sampleKernel", null);
for(int i = 0; i < memObjects.length; i++) {
clSetKernelArg(kernel, i, Sizeof.cl_mem, Pointer.to(memObjects[i]));
}
cl_event event = new cl_event();
long global_work_size[] = new long[]{ARRAY_SIZE};
long local_work_size[] = new long[]{1};
long start = System.nanoTime();
clEnqueueNDRangeKernel(commandQueue, kernel, 1, null,
global_work_size, local_work_size, 0, null, event);
clWaitForEvents(1, new cl_event[] {event});
long end = System.nanoTime();
System.out.println("Information for " + getDeviceInfoString(device, CL_DEVICE_NAME));
System.out.println("\tGPU Runtime: " + getRuntime(event));
System.out.println("\tJava Runtime: " + ((end - start) / 1e6) + " ms");
clReleaseEvent(event);
for(int i = 0; i < memObjects.length; i++) {
clReleaseMemObject(memObjects[i]);
}
clReleaseKernel(kernel);
clReleaseProgram(program);
clReleaseCommandQueue(commandQueue);
clReleaseContext(context);
}
float[] result1 = new float[ARRAY_SIZE];
float[] result2 = new float[ARRAY_SIZE];
float[] result3 = new float[ARRAY_SIZE];
float[] result4 = new float[ARRAY_SIZE];
long start = System.nanoTime();
for(int i = 0; i < ARRAY_SIZE; i++) {
result1[i] = sourceA[i] + sourceB[i];
result2[i] = sourceA[i] - sourceB[i];
result3[i] = sourceA[i] * sourceB[i];
result4[i] = sourceA[i] / sourceB[i];
}
long end = System.nanoTime();
System.out.println("JVM Benchmark: " + ((end - start) / 1e6) + " ms");
}
private static String getDeviceInfoString(cl_device_id device, int parameter) {
long[] bufferLength = new long[1];
clGetDeviceInfo(device, parameter, 0, null, bufferLength);
byte[] buffer = new byte[(int) bufferLength[0]];
clGetDeviceInfo(device, parameter, bufferLength[0], Pointer.to(buffer), null);
return new String(buffer, 0, buffer.length - 1);
}
private static String getRuntime(cl_event event) {
long[] start = new long[1];
long[] end = new long[1];
clGetEventProfilingInfo(event, CL_PROFILING_COMMAND_START, Sizeof.cl_ulong, Pointer.to(start), null);
clGetEventProfilingInfo(event, CL_PROFILING_COMMAND_END, Sizeof.cl_ulong, Pointer.to(end), null);
long nanos = end[0] - start[0];
double millis = nanos / 1e6;
return millis + " ms";
}
}
Выход:
Information for Intel(R) Core(TM) i7-6820HQ CPU @ 2.70GHz
GPU Runtime: 639.986906 ms
Java Runtime: 641.590764 ms
Information for Quadro M2000M
GPU Runtime: 794.972 ms
Java Runtime: 1191.357248 ms
Information for Intel(R) HD Graphics 530
GPU Runtime: 1897.876624 ms
Java Runtime: 2065.011125 ms
JVM Benchmark: 192.680669 ms
Похоже, это указывает на то, что более мощная карта NVIDIA на самом деле работает лучше, чем Intel, как я и ожидал. Но...
- Почему процессор все еще быстрее?
- Почему нормальная Java внезапно оказывается намного быстрее?
1 ответ
Я все еще копаюсь и пытаюсь понять это, но я начну публиковать фактический ответ здесь, чтобы принести пользу любым другим невежественным новичкам, таким как я. Надеюсь, кто-то, кто менее невежественен, скоро придет, чтобы исправить меня во всем, в чем я не прав, но, по крайней мере, эти другие невежественные новички смогут увидеть, через что я работал, и извлечь из этого уроки.
Как я отметил в редактировании вопроса, часть странных результатов была связана с тем, что я полагался на Java, чтобы сказать мне, как быстро все работает. Это не совсем неправильно, я думаю, но я неправильно понял данные. Время выполнения Java будет включать в себя время, которое требуется Java для перевода всего в память GPU и из нее, тогда как время выполнения OpenCL просто сообщит, сколько времени потребуется для запуска; в конце концов, OpenCL на самом деле не знает и не заботится о том, как это называется. Включение профилирования OpenCL и использование событий для отслеживания времени выполнения помогло мне это прояснить. Это также объясняет очень маленький разрыв между временами выполнения для ЦП; на самом деле это не было переключение устройств, поэтому передача памяти не происходила.
Я также заметил, что код, который я имел выше, имеет серьезный недостаток. При постановке в очередь команды ядра CL.clEnqueueNDRangeKernel принимает девять аргументов. Шестой аргумент называется "local_work_size"; это указывает на количество "рабочих групп", которые вы хотите, чтобы OpenCL использовал для запуска вашего кода. Самым близким аналогом, который я могу придумать для Java, являются потоки; больше потоков (обычно) означает, что больше работы может быть сделано за один раз (до определенного момента). В приведенном выше коде я делал то, что показал пример, и велел OpenCL использовать одну рабочую группу; в основном, чтобы запустить все в одном потоке. Насколько я понимаю, это именно НЕПРАВИЛЬНАЯ вещь в GPGPU; весь смысл использования графического процессора заключается в том, что он может обрабатывать гораздо больше вычислений за раз, чем процессор. Принуждение графического процессора к выполнению одного вычисления за раз побеждает точку. Похоже, что лучший подход здесь - просто оставить этот шестой аргумент пустым; это дает указание OpenCL создать столько рабочих групп, сколько считает нужным. Вы можете указать число, но максимально допустимое количество зависит от вашего устройства (вы можете использовать CL.clGetDeviceInfo, чтобы получить атрибут CL_DEVICE_MAX_WORK_GROUP_SIZE вашего устройства для определения абсолютного максимума, но это становится более сложным, если вы используете более одного измерения),
Короткая версия:
- Профилирование OpenCL даст вам лучшую статистику синхронизации, чем Java (однако использование обоих поможет показать задержку, необходимую для переключения между CPU и GPU)
- Не указывайте local_work_size при вызове CL.clEnqueueNDRangeKernel - это позволяет OpenCL автоматически обрабатывать "многопоточность"
Новые результаты:
Information for Quadro M2000M
GPU Runtime: 35.88192 ms
Java Runtime: 438.165651 ms
Information for Intel(R) Core(TM) i7-6820HQ CPU @ 2.70GHz
GPU Runtime: 166.278112 ms
Java Runtime: 167.128259 ms
Information for Intel(R) HD Graphics 530
GPU Runtime: 90.985728 ms
Java Runtime: 239.230354 ms
JVM Benchmark: 177.824372 ms