Тяжелый старт
Всем привет! Какое-то время назад я начал копать тему с OpenCL под C#. Но наткнулся на трудности, связанные с тем, что не то, что под C#, а вообще по этой теме очень мало материала. Какую-то вводную по OpenCL можно почерпнуть здесь. Так же простой, но работающей старт OpenCL описан вот тут. Ни на йоту не хочу обидеть авторов, но все статьи, что я находил на русском (и на хабре в том числе) страдают одной и той же проблемой — очень мало примеров. Документация есть, её много и как принято для хорошей документации читается сложно. В своей статье (а если всё будет нормально, то и в цикле статей), я постараюсь поподробней описать эту область, с точки зрения человека, который начал её копать с нуля. Думаю такой подход будет полезен тем кто хочет быстро стартовать в высоко производительных вычислениях.
Первоначально я хотел написать статью-минисамоучитель OpenCL, которая содержала в себе информацию, о там что это, как устроено, как писать код и какие-то рекомендации, основанные на моем опыте. Но в процессе понял, что если даже быть кратким, то уткнусь в ограничения объема статьи. Потому что, имхо, статья должна быть такого объема, чтобы усвоить её объем было не сложно. По этому в данной статье (которая станет первой) я планирую описать, то как стартовать в OpenCL, проверить что локально все корректно законфигурино, и написать простейшую программу. Вопросы с устройством памяти, архитектурой и прочим будут описаны в следующих статьях.
Друзья, сразу хотел бы сказать, что мой опыт в OpenCL пока, к сожалению, далек от гуру/йода уровня, но на вопросы постараюсь отвечать изо всех сил. А если что-то не знаю, то буду делиться ресурсами и видением того как это на самом деле должно работать.
Что. Где. Как.
OpenCL это технология связанная с параллельными компьютерными вычислениями на различных типах графических и центральных процессоров. Тема с параллельным вычислениями на GPU совсем недавно широко продвигалась вместе с технологией CUDA. Данное продвижение в основном обеспечивалось усилиями компании Nvidia. Отличия OpenGL и CUDA уже широко обсуждались.
OpenСL позволяет работать как с CPU так и с GPU, но думаю нам более интересно будет сосредоточиться на работе с GPU. Для использования данной технологии понадобиться мало мальски современная видеокарта. Главное это проверить, что устройство функционирует нормально. На всякий случай напоминаю что это можно сделать в диспетчере устройств.
Если в данном окне вы видите какие-то фейлы или ворнинги, то вам прямая дорога на сайт производителя вашей видеокарты. Свежие драйвера должны решить проблему с функционированием железа и как следствие дать доступ к мощностям OpenCL.
Первоначально я планировал использовать OpenCL по C#. Но наткнулся на проблему, что все существующие фреймворки типа Cloo или OpenCLNet являются самописными и Khronos не имеет к ним никакого отношения и следовательно не гарантирует их стабильную работу. Ну и все мы помним главную проблему — очень мало примеров. Исходя из этого вначале я бы хотел представить примеры написанные на C++, а уже потом, получив подтверждение того что OpenCL ведет себя так как мы ожидаем, привинтить прокси в виде C# фреймворка.
Итак, чтобы использовать OpenCL через С++ необходимо найти его API. Для этого открывайте переменные среды, и ищете там переменную со страшным названием, намекающую своим названием на производителя вашей видеокарты. У меня данная переменная называется «AMDAPPSDKROOT». После этого можете посмотреть что лежит по указанному пути. Там ищите папочку include\CL.
Кстати, обычно в папочки include, рядом с папкой CL лежит и папка GL, предоставляющая доступ к знаменитой графической библиотеки.
Теперь создаем проект в Visual Studio, подключаем в свойствах проекта папку include (в моем случае $(AMDAPPSDKROOT)\include\) и в бой!
Инфраструктура
Мы должны помнить, что мы будем работать с OpenCL не через API, а при помощи API. Вроде бы кажется, что эти две фразы практически идентичны, но это не так. К примеру, вспомните OpenGL. Как происходит там работа (урощённый вариант) — вначале мы настраиваем какие-то общие параметры, а потом прямо из кода вызываем методы типа «нарисовать сферу», «изменить параметры источника света» и т.д.
Так вот в OpenCL сценарий другой:
- При помощи API получаем доступ к устройствам, которые поддерживают OpenCL. Это часть приложения обычно называется хостом;
- Пишем код который будет выполняться на устройстве. Этот код называется kernel. Данный код о хосте вообще ничего не знает. Его может дернуть любой хост
- При помощи API прогружаем код kernel и запускаем его выполнение на выбранном устройстве.
Как видите наше приложение будет иметь комплексную инфраструктуру. Давайте займемся ее настройкой!
Так как на предыдущем шаге мы предусмотрительно подсоединили папочку include, то теперь вы можем просто добавить ссылку на заголовочный файл cl.h, который даст доступ к API. При добавление cl.h, стоит добавить проверку выбора платформы:
#ifdef __APPLE__
#include <OpenCL/opencl.h>
#else
#include <CL/cl.h>
#endif
Теперь необходимо выбрать устройство на котором будет отрабатывать наш код и создать контекст в котором будут жить наши переменные. Как это сделать показано ниже:
/* получить доступные платформы */
ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms);
/* получить доступные устройства */
ret = clGetDeviceIDs(platform_id, CL_DEVICE_TYPE_DEFAULT, 1, &device_id, &ret_num_devices);
/* создать контекст */
context = clCreateContext(NULL, 1, &device_id, NULL, NULL, &ret);
/* создаем команду */
command_queue = clCreateCommandQueue(context, device_id, 0, &ret);
Обращаю внимание на переменную ret. Это переменная, которая содержит числовое значение которое возвращает та или иная функция. Если ret== 0, то функция выполнилась корректно, если нет, то значит произошла ошибка.
Так же заслуживает внимание константа CL_DEVICE_TYPE_DEFAULT, она запрашивает устройство, которое используется для вычислений на OpenCL по умолчанию. Вместо данной константы могут быть использованы другие. К примеру:
- CL_DEVICE_TYPE_CPU — запросит существующие CPU.
- CL_DEVICE_TYPE_GPU — запросит существующие GPU.
Kernel
Отлично. Настроили инфраструктуру. Теперь возьмемся за kernel. Kernel — это просто функция объявление, которой начинается с ключевого слова __kernel. Синтаксис языка программирования OpenCL базируется на стандарте C99, но имеет ряд специфических и очень важных изменений. Об этом будет (я очень надеюсь) отдельная статья. Пока базовая информация:
- Код который, будет дергаться с хостовой части, для исполнения, должен начинаться с ключевого слова __kernel;
- Функция с ключевым словом __kernel всегда возвращает void;
- Существуют квалификаторы типов памяти: __global, __local, __constant, __private, которые будут определять, в какой памяти будут храниться переменные. Если квалификатора перед переменной нет, то она является __private;
- «Общение» между хостом и kernel будет через параметры kernel. Чтобы kernel мог что-то передать хосту через параметр, параметр должен быть с квалификатором __global (пока будем использовать только __global);
- Код kernel принято хранить в файле с расширением cl. Но по сути подобный код может генерироваться и на лету. Это позволяет обойти некоторые ограничения. Но об этом в другой раз :)
Простейший пример kernel приведен ниже:
__kernel void test(__global int* message)
{
// получаем текущий id.
int gid = get_global_id(0);
message[gid] += gid;
}
Что делает данный код. Первое — получает глобальный id work-item который сейчас выполняется. Work-item — это то что и выполняет наш kernel. Так как мы имеем дела с параллельными вычислениями, то для каждого work-item создается свой kernel который ничего не знает о других. И никто не может гарантировать в каком порядке все work-item отработают. Но об этом подробней будет в отдельной статье (уже утал это повторять). В нашем примере это по сути индекс элемента в массиве, потому что мы будем каждый элемент массива обрабатывать в отдельном work-item. Думаю вторую строчку строчку в kernel комментировать излишни :)
Формируем kernel
Следующий шаг скомпилировать, то что лежит в файле *.cl. Делается это следующим образом:
cl_program program = NULL;
cl_kernel kernel = NULL;
FILE *fp;
const char fileName[] = "../forTest.cl";
size_t source_size;
char *source_str;
int i;
try {
fp = fopen(fileName, "r");
if (!fp) {
fprintf(stderr, "Failed to load kernel.\n");
exit(1);
}
source_str = (char *)malloc(MAX_SOURCE_SIZE);
source_size = fread(source_str, 1, MAX_SOURCE_SIZE, fp);
fclose(fp);
}
catch (int a) {
printf("%f", a);
}
/* создать бинарник из кода программы */
program = clCreateProgramWithSource(context, 1, (const char **)&source_str, (const size_t *)&source_size, &ret);
/* скомпилировать программу */
ret = clBuildProgram(program, 1, &device_id, NULL, NULL, NULL);
/* создать кернел */
kernel = clCreateKernel(program, "test", &ret);
Типы cl_program и cl_kernel определены в cl.h. Сам сценарий довольно прост — загружаем файл, создаем бинарник (clCreateProgramWithSource) и компилируем. Если переменная ret по прежнему содержит 0, то вы все сделали правильно. И останется только создать сам kernel. Важно, чтобы имя передаваемое в команду clCreateKernel, совпадало с именем kernel в файле cl. В нашем случае это «test».
Параметры
Я уже упоминал, что «общения» kernel с хостом происходит за счет записи/чтения в параметры, которые передаются в kernel. В нашем случае это параметр message. Параметры, которые позволяют вот так общаться хосту с kernel, называются буферами(buffer). Давайте создадим такой буфер на стороне хоста и передадим в kernel через API:
cl_mem memobj = NULL;
int memLenth = 10;
cl_int* mem = (cl_int *)malloc(sizeof(cl_int) * memLenth);
/* создать буфер */
memobj = clCreateBuffer(context, CL_MEM_READ_WRITE, memLenth * sizeof(cl_int), NULL, &ret);
/* записать данные в буфер */
ret = clEnqueueWriteBuffer(command_queue, memobj, CL_TRUE, 0, memLenth * sizeof(cl_int), mem, 0, NULL, NULL);
/* устанавливаем параметр */
ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&memobj);
Важно отметить константу CL_MEM_READ_WRITE, она означает, что мы у нас есть права для буфера на чтение и запись, на стороне kernel. Так же могут быть использованы константы типа CL_MEM_WRITE_ONLY, CL_MEM_READ_ONLY и др. Так же в методе clSetKernelArg, важен второй аргумент, он содержит индекс параметра. В данном случае 0, так как параметр message идет первым в сигнатуре kernel. Если бы он шел вторым, то мы бы написали:
/* устанавливаем параметр */
ret = clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *)&memobj);
clEnqueueWriteBuffer записывает данные из массива mem в буфер memobj.
Ну что в целом все готово. Осталось только выполнить kernel.
Исполняем kernel
Погнали, отправляем код на исполнение:
size_t global_work_size[1] = { 10 };
/* выполнить кернел */
ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL, global_work_size, NULL, 0, NULL, NULL);
/* считать данные из буфера */
ret = clEnqueueReadBuffer(command_queue, memobj, CL_TRUE, 0, memLenth * sizeof(float), mem, 0, NULL, NULL);
global_work_size содержит число work-item которые будут созданы. Я уже говорил, что на обработку каждого элемента массива у нас будет свой work-item. Элементов в массиве у нас 10, следовательно work-item содержит 10. clEnqueueNDRangeKernel особых вопросов порождать не должна — просто запускает указанный kernel заданное число раз. clEnqueueReadBuffer считывает данные из буфера с именем memobj и помещает данные в массив mem. Данные в mem и есть наш результат!
Итоги и выводы
Друзья, вот так я представляю старт в OpenCL для новичка. Надеюсь на ваши конструктивные замечания в комментариях, чтобы можно было внести апдейты в будущем. Я пытался быть кратким, но все равно объем вышел не маленький. Так что могу сказать, что материала для 2-3 статей найти еще смогу.
Спасибо, всем кто дочитал до конца!
Комментарии (32)
alexac
05.07.2015 09:04Пожалуйста уберите try/catch вокруг чтения файла. Он там бесполезен. Вы используете только функции libc, которые не бросают исключения. Единственная ошибка, которую в этом коде еще можно было бы получить — ошибка выделения памяти, что в текущем виде все равно приведет к segmentation fault, а не к исключению. Тот кусок стоило написать как-то так:
const char fileName[] = "../forTest.cl"; size_t source_size; std::unique_ptr<char[]> source; try { std::ifstream stream; stream.exceptions(std::ios_base::bad_bit | std::ios_base::fail_bit | std::iOS_base::eof_bit); stream.open(fileName, std::ios_base::in); stream.seekg(0, std::ios_base::end); size_t source_size = stream.tellg(); stream.seekg(0, std::ios_base::beg); source.reset(new char[source_size]); stream.read(source.get(), source_size); } catch (const std::exception& e) { std::cerr << "Can't read kernel: " << e.what() << std::endl; exit(1); }
vilgeforce
05.07.2015 12:32«Главное это проверить, что устройство функционирует нормально.» — ни разу не гарантия, что у вас будут нормально перечисляться устройства, как показывает практика.
clGetPlatformIDs у вас вызывается, но нет выделения памяти под результаты и освобождения ее потом. И такая же проблема с другими функциями.
rznELVIS Автор
05.07.2015 12:57Тестировал на 5 компьютерах/ноутах. Если драйвера стоят актуальные то проблем не бывало. Слышал о возможных проблемах с устройствами, но сам не встречал. А хотелось писать о том что видел сам.
Про освобождение ресурсов разумеется поговорим. Все это будет описано. Тут уже еле влезал в объем. Сейчас описан простой вариант, который ресурсы сильно не ест. ресурсы разумеется освобождаются.
stepik777
05.07.2015 12:36+1К примеру, вспомните OpenGL. Как происходит там работа (урощённый вариант) — вначале мы настраиваем какие-то общие параметры, а потом прямо из кода вызываем методы типа «нарисовать сферу», «изменить параметры источника света» и т.д.
Это было в старом OpenGL, но и там можно было шейдеры писать, а начиная с OpenGL 3 (которому уже почти 7 лет) остались только шейдеры.
Так вот в OpenCL сценарий другой:rznELVIS Автор
05.07.2015 13:00Спасибо. приму к сведению. Работал только со старым OpenGL. По-этому и вспомнил его для хорошей метафоры. к тому же совсем недавно тыкал OpenGL через библиотеку Tao (студентам показывал), в шейдеры не лазил.
deniskreshikhin
05.07.2015 13:02Да, очень верное замечание. Ведь kernel'ы как раз и выполняются на шейдерных процессорах.
stgatilov
05.07.2015 13:35Я думаю, при освоении OpenCL лучше перейти на официальные C++-обёртки как можно быстрее. Использование C-шного API — лишняя трата ценного времени: тонны параметров, void*, обработка ошибок и пр.
P.S. Khronos как-то не так написан…rznELVIS Автор
05.07.2015 22:50Буду рад ссылочкам, на понравившиеся вам фреймворки.
Khronos поправил)) спасибоstgatilov
06.07.2015 07:30+1Это не фреймворк, это именно обёртки над C API, официальные от Khronos.
Все команды в обёртках являются эквивалентами команд C-шного API, и наоборот, для каждой команды C-шного API есть эквивалент в обёртках. Параметры почти такие же, названия примерно одинаковые. То есть отдёльно «учить» C++ обёртки не нужно: зная C API, можно легко начать пользоваться обёртками.
Среди бонусов:
- всюду нормальная типизация вместо void * и размеров в байтах
- использование std::vector или cl::vector, cl::string вместо char*
- автоматическое отслеживание времени жизни объектов (типа умных указателей)
- автоматическая обработка ошибок через исключения (нужно включить)
Хедер
Спецификация
Пока есть только для OpenCL 1.2, новые версии запаздывают.
BelBES
06.07.2015 11:39А есть какие-то бенчмарки на тему того, насколько производительность отличается у кода нормально написанного через ANSI C API и при помощи оберток? Ведь скорей всего из оберток генерируется менее оптимальный код.
stgatilov
06.07.2015 12:24В спецификации (ссылка выше) об этом написано: «The C++ API corresponds closely to the underlying C API and introduces no additional execution overhead».
Вопрос про производительность вряд ли возникнет у человека, который использует C++ wrappers и мельком видел их внутренности.
Вариант API никак не влияет на код собственно вычислительных программ OpenCL, так что ядра работают абсолютно с одинаковой скоростью. Копирование данных запускает те же самые команды C-шного API, то есть работает столько же времени. Все понятия, такие как асинхронные вызовы например, одинаковы и в C, и в C++.
Чаще всего разница только в том, что вместо прямого вызова C-шной функции вызывается метод класса C++, который внутри вызывает ту же C-шную функцию. Скорее всего он инлайнится, а даже если и нет, overhead от лишнего вызова будет ничтожно мал по сравнению с временем работы самого вызова в драйвере. Ну может быть какой-то overhead есть из-за того, что используется std::vector<cl::device> и std::vector<std::string> вместо вручную выделенных кусков памяти. Но это опять же ничтожно малое время по сравнению с временем инициализации девайса и компиляции программы. Обработка ошибок тоже может отнимать время, но чтобы это стало заметным, надо написать невероятно неэффективную программу с огромным количеством API-вызовов.
По сути C++ обёртка — это просто более удобное использование C-шного API.
deniskreshikhin
06.07.2015 12:36В OpenCL/OpenGL не очень важно как происходят вызовы. напрямую или нет, данные же передаются через буффер. Если правильно использовать API то может быть 1 вызов на миллионы и миллиарды операций GPU. Здесь даже интерпретируемые языки показывают сносное быстродействие.
Так что основные затраты идут на подготовку данных, а не на вызовы, но здесь между C/C++ нет особой разницы.
RomanArzumanyan
06.07.2015 12:37API написан на С неспроста. Как только начнёте оптимизировать производительность (использовать события и их списки, разбивать обьекты памяти для одновременного использования, и т. п.), наверняка вернётесь к низкоуровневому API. ИМХО, официальные С++ обёртки добавляют удобства только в момент инициализации.
stgatilov
06.07.2015 13:17+1Мне кажется, API написан на C в первую очередь потому, что его больше не на чем писать. C-шные функции легко использовать из любого языка программирования в любой ситуации. Если делать C++ API первостепенным, то начнётся куча проблем (радости с именованием символов, версии и настройки STL). К тому же, драйверы как-то принято писать на чистом C, а не на C++ со всякими std::vector и std::string=)
Конечно, мне пока что не были необходимы ни синхронизация по событиям, ни тем более разделение буфера на части. Однако мне кажется, что как минимум синхронизация по событиям в C++ обёртках есть, и нет проблем её использовать. В любом случае, такие проблемы уже совсем не для новичков.
К тому же, никто не мешает использовать C++ обёртки почти всюду, а в случае возникновения трудностей перейти на C API в одном конкретном проблемном месте. В этом ещё одно отличиё обёрток от фреймворка: обёртки и исходный API можно легко смешивать (с фреймворкам дело обычно сложнее).
Например, мне пришлось писать извлечение бинарного кода программы на C API из-за бага обёрток. Надо сказать, получилось 25 уродливых строчек с кучей звёздочек, sizeof, и самопальных макросов SAFE_CALL. Если бы не баг в обёртках (который, кстати, никто не мешает исправить самому локально), было бы строчек 2-5.RomanArzumanyan
06.07.2015 15:18C-шные функции легко использовать из любого языка программирования в любой ситуации
С этим не поспоришь.
Пытался пару раз использовать С++ обёртку — мне не понравилось то, что она ничуть не более «интеллектуальная», чем оригинал: например, при чтении буффера нужно всё так же указавать размер, смещение и т. п. — как будто бы класс объекта памяти не должен инкапсулировать в себе эти данные. А если уж указываешь (почти) всё руками, то, ИМХО, невелика разница — делать это из С или из плюсов.stgatilov
06.07.2015 15:51Да, C++ обёртка не решает волшебным образом все проблемы именно потому, что она почти полностью соответствует C-шному API.
Чтобы не мучиться постоянно, придётся пописать какие-то классы поверх неё (например типизированный буфер). Будет проще, чем поверх C-шного API.
Salabar
06.07.2015 16:30По-моему, базовый ОпенЦЛ это аналог JDBC в СУБД. Писать программы но можно, но в определенный момент есть риск немножко поехать, так что для начала следует написать себе небольшой фреймворк под задачу. То, что можно писать на один параметр благодаря плюсам это классно, но это не избавит от десятков clSetKernelArg подряд. С другой стороны, если в качестве будущего OpenCL разработки собираются позиционировать этот ад: www.khronos.org/registry/sycl/specs/sycl-1.2.pdf, то это даже не такое и зло.
stgatilov
06.07.2015 17:51но это не избавит от десятков clSetKernelArg подряд
Десяток clSetKernelArg подряд — это C API.
В C++ обёртках ядро принято вызывать примерно как обычную функцию.
Вот пример:
//устанавливаем параметры запуска (кол-во потоков, очередь) auto launcher = kernel.bind(queue, AlignUp(paramsCnt, 64), 64); //вызываем ядро, передавая параметры как в обычную функцию launcher(paramsBuffer, paramsCnt, resultsBuffer);
Здесь аргументы — объекты типа cl::Buffer и примитивные типы.
Хотя должен заметить, что обёртки писались под старый стандарт C++, поэтому количество аргументов ограничено 32 (хотя вроде не сложно это число увеличить). Бывало, в это ограничение я не укладывался =)Salabar
07.07.2015 11:14Так и с С API никто не мешает функцию написать. Суть в том, что OpenCL сам по себе мало помогает определять зависимости между вычислениями. Низкоуровневый слишком. Это даже хорошо, можно даже пойти еще дальше и сделать что-то типа Vulkan с поддержкой OpenCL C, чтобы не было всяких нубских clCreateBuffer, но если я в курсаче на десяток ядер запутался, то как что-то серьезное на этом писать я не представляю. Я хочу сесть и подумать над либой, которая позволит составить граф вычислений декларативно, чтобы библиотека сама расставляла примитивы синхронизации, создавала промежуточные буферы, но, честно говоря, даже не представляю, с чего начинать.
stgatilov
07.07.2015 11:59Во-первых, можно для начала написать прототип с синхронными запусками ядер и копирований. В большинстве случаев будет прекрасно работать (если ядра делают много вычислений).
Потом можно попробовать что-то типа ленивых вычислений. Достаточно реализовать класс future-буфера, у которого внутри есть cl::Buffer (где будут данные) и cl::Event (когда данные в буфере будут правильными). Ну а дальше сделать обёртку над каждой командой, которая будет сама устанавливать списки ожидания для команд и записывать куда следует cl::Event готовности результата.
В целом кажется, что стоит покурить ключевые слова вроде future/promise/async.
P.S. Писать отдельное декларативное описание ядер и зависимостей между ними в каком-то особом формате — кажется излишним.
stgatilov
06.07.2015 18:25С другой стороны, если в качестве будущего OpenCL разработки собираются позиционировать этот ад: www.khronos.org/registry/sycl/specs/sycl-1.2.pdf, то это даже не такое и зло.
SYCL — замечательная задумка. Пока что она сильно молодая, чтобы всерьёз возлагать на неё надежды. Судя по всему, это будет платформенно-независимой версией CUDA runtime.
Должен заметить, что CUDA runtime на порядок удобнее в использовании, чем OpenCL, что затрудняет переход CUDA-поклонников на OpenCL (особенно учёных, которым чихать на кроссплатформенность).
Конечно, SYCL выглядит как-то далековато от простого OpenCL, всюду куча абстракций, на первый взгляд ненужных. Однако какие-то вещи оттуда существенно облегчают работу.
Например, я реализовал в нашем проекте «двойной» типизированный буфер, который существует одновременно на host и device, и синхронизируется автоматически прозрачно для пользователя. Естественно, пользователь должен корректно вызывать функции доступа к сторонам буфера, указывая каждый раз: нужен доступ «только на чтение» или «на чтение и запись».
Чуть позже я узнал о SYCL. В нём sycl::buffer является ровно тем же самым общим типизированным буфером, и для работы с ним надо ровно так же вызывать функцию с указанием типа доступа.
Также в SYCL есть более простой способ использования асинхронной работы.
Короче говоря, лично я надеюсь, что в будущем смогу писать на SYCL =)
igor_suhorukov
07.07.2015 19:20А как происходит отладка OpenCL kernel функций и C#+host api?
Salabar
07.07.2015 20:04+1На некоторых платформах можно из ядер вызывать printf. Для всего остального придется сначала отладить их как обычные Си-функции и надеяться, что багов с многопоточностью там не осталось.
RomanArzumanyan
07.07.2015 21:42+1Если вы пишете под AMD, то можно отлаживать код ядер из AMD CodeXL. Intel тоже в каком-то виде поддерживает отладку внутри ядра. В остальных случаях — уповать на наличие printf и/или сливать результаты вычислений в отладочные объекты памяти и делать postmortem.
ProPupil
Для c# есть целых две библиотеки:Cloo и OpenClTemplate. Как раз в ваших приведённых ссылках есть ссылка на статью, посвящённую этим двум библиотекам, нос оглашусь с Вами, что на русском все-таки маловато, хоть нагло язычная документация и достаточно чёткая. По крайней мере, у меня не возникало проблем с её пониманием.
rznELVIS Автор
тут для меня проблема по другому чуть выглядит. когда осваиваешь OpenCL через подобную Net-библиотеку, то иногда начинаешь сомневаться в чем проблема (при наличие ошибки) в библиотеки или в коде на OpenCL. а в в работе с OpenCL ошибки можно получить даже при корректном kernel. Пример некорректное значение global и local work-item size. И по-этом чтобы не бояться эффекта «сломанного телефона», то я и предлагаю получить подтверждение очевидного, поработав с OpenCL через родной API, а потом уж гоу и в .Net.
HomoLuden
aka «Дырявая» или «Утекающая Абстракция»