- PVSM.RU - https://www.pvsm.ru -
Надеюсь, при прочтении этого блока в своём ридере, моя картинка вас не напугала. Но сегодня, я хочу описать применение взаимодействия технологии CUDA с OpenGL на примере моего небольшого pet-примера, первую версию которого я описывал в статье [2] ранее. Тех, кому интересен раздел, известный под английским названием CUDA and OpenGL interoperability, прошу кликать сюда.
Все статьи в Гугле рассказывают о теории следующим образом: по классике, перед рендерингом кадра, программа формирует логику и инициализацию сцены в CPU, и потом уже GPU занимается её рендерингом. А теперь, представьте, что вы ту же саму сцену инициализируете в многопоточной среде. Например, вы можете сгенерировать массив точек в трёхмерной системе координат с помощью CUDA, или же вообще нарисовать картинку самостоятельно в своём ядре (kernel), и с помощью OpenGL просто отобразить результат на экране.
Вот этим я и собираюсь заняться на небольшом примере. Ах да, если хотите прочесть больше слов о теории, чем я описал здесь, можете пробежаться глазами здесь [3]. В моей задаче, я хочу посчитать значение напряжённости электрического поля в каждой точке текущего кадра и сопоставить ей некоторый цвет. И чем быстрей будет проходить этот расчёт, тем больше кадров в секунду я смогу сгенерировать.
Когда начинаешь создавать шаблон проекта, очень важно не упустить ряд моментов:
#include <GL/glew.h>
#include <GL/freeglut.h>
...
#include <cuda_gl_interop.h>
p.s. если что, вас предупредит компилятор.
GLuint vbo;
struct cudaGraphicsResource *cuda_vbo_resource;
void createVBO(GLuint *vbo, struct cudaGraphicsResource **vbo_res, unsigned int vbo_res_flags) {
unsigned int size = width * height * sizeof(uchar4);
glGenBuffers(1, vbo);
glBindBuffer(GL_PIXEL_UNPACK_BUFFER_ARB, *vbo);
glBufferData(GL_PIXEL_UNPACK_BUFFER_ARB, size, NULL, GL_DYNAMIC_DRAW);
HANDLE_ERROR( cudaGraphicsGLRegisterBuffer(vbo_res, *vbo, vbo_res_flags) );
}
void deleteVBO(GLuint *vbo, struct cudaGraphicsResource *vbo_res) {
HANDLE_ERROR( cudaGraphicsUnregisterResource(cuda_vbo_resource) );
glBindBuffer(GL_PIXEL_UNPACK_BUFFER_ARB, *vbo);
glDeleteBuffers(1, vbo);
*vbo = 0;
}
Основная идея в том, что позже, в коде, можно получить указатель на созданный буфер в виде массива пикселей, и потом менять его в ядре. К сожалению, я не сильно сведущ в OpenGL, и потом мне не ясен момент, почему часто перед удалением, вызывают функцию glBindBuffer [4] ещё раз, и передают 1 в качестве первого аргумента. Если я правильно понимаю, то если я использую только один буфер в своём приложении, то я сразу могу очистить его после завершения.
HANDLE_ERROR( cudaGLSetGLDevice(deviceId) );
...
createVBO(&vbo, &cuda_vbo_resource, cudaGraphicsMapFlagsWriteDiscard);
...
deleteVBO(&vbo, cuda_vbo_resource);
Вроде бы всё! Теперь мы можем обработать idle событие у приложения, где заниматься формированием кадра, а потом просто отрисовывать его. Следует напомнить, что мы рисуем кадр сразу же в буфере видеокарты, и потом просто просим её его отобразить.
void idle(void) {
uchar4* dev_screen;
size_t size;
HANDLE_ERROR( cudaGraphicsMapResources(1, &cuda_vbo_resource, 0) );
HANDLE_ERROR( cudaGraphicsResourceGetMappedPointer((void**) &dev_screen, &size, cuda_vbo_resource) );
// Render Image
renderFrame<<<blocks, threads>>>(dev_screen);
HANDLE_ERROR( cudaDeviceSynchronize() );
HANDLE_ERROR( cudaGraphicsUnmapResources(1, &cuda_vbo_resource, 0) );
glutPostRedisplay();
}
p.s. Из кода я выкинул участок, отвечающий за вычисление времени создания кадра и отображение его в заголовке приложения.
Когда я отошел от использования простых структур представляющих заряд, в сторону float4, то смог сделать выполнение кода быстрее. Так, например, кадр 1024*768 с 10ю зарядами создаётся теперь за 25-30 мс. И это позволило мне создавать плавную анимацию с помощью мыши — drag & drop:
Ну и как всегда, готовый код вы можете найти здесь [5]. И если у вас возникнут вопрос или пожелания, пишите мне. Постараюсь исправиться.
В будущем планирую сделать режим отображающий напряжённость поля в виде опилок выброшенных на бумагу. Идея — рисовать короткие отрезки через равные промежутки. Ну и если уж совсем повезет, то попробую реализовать это в 3д :)
Автор: Vest
Источник [6]
Сайт-источник PVSM.RU: https://www.pvsm.ru
Путь до страницы источника: https://www.pvsm.ru/opengl/27878
Ссылки в тексте:
[1] Image: http://habrastorage.org/storage2/115/b80/47a/115b8047ac4a068e1b808534ecc95b2e.png
[2] статье: http://habrahabr.ru/post/164499/
[3] здесь: http://www.dyn-lab.com/articles/cl-gl.html
[4] glBindBuffer: http://www.opengl.org/sdk/docs/man/xhtml/glBindBuffer.xml
[5] здесь: https://github.com/Vest/ElectroField
[6] Источник: http://habrahabr.ru/post/170461/
Нажмите здесь для печати.