Всем привет,Пишу код на Ubuntu 14, Tegra 3, камера с разрешением 4224x3156, yuv422, 26Мб. Забираю кадр каждые 70мс, далее memcpy в pinned memory host, cudaMemcpy из pinned host в divice memory и далее обработка в GPU.
Memcpy занимает 200мс. Это больше чем частота кадров. Как можно ускорить загрузку буфера кадра в память GPU?
Можно ли использовать v4l2 IO_METHOD_USERPTR в качестве pinned memory, что бы не тратить время на memcpy.
С уважением Виктор.
> Всем привет,...
> Можно ли использовать v4l2 IO_METHOD_USERPTR в качестве pinned memory, что бы не
> тратить время на memcpy.
> С уважением Виктор.шаг 1.
https://linuxtv.org/downloads/v4l-dvb-apis/
&
https://linuxtv.org/downloads/v4l-dvb-apis/io.htmlшаг 2.
смотрим в /usr/include/*
что-то вроде video*.h на предмет V4L2_MEMORY_USERPTR
Спасибо за ответ.
Делал по примеру отсюда: https://linuxtv.org/downloads/v4l-dvb-apis/capture-example.htmlМетод MMAP:
В process_image всего одна строка memcpy( pinned, p, size ) //200ms
Инициализирую так: cudaMallocHost( pinned, size )
Метод USERPTR:
Пробовал вместо: buffers[n_buffers].start = malloc(buffer_size)
делать: cudaMallocHost( buffers[n_buffers].start, buffer_size )Прога зависает здесь:
case IO_METHOD_USERPTR:
for (i = 0; i < n_buffers; ++i) {
struct v4l2_buffer buf;CLEAR(buf);
buf.type = V4L2_BUF_TYPE_VIDEO_CAPTURE;
buf.memory = V4L2_MEMORY_USERPTR;
buf.index = i;
buf.m.userptr = (unsigned long)buffers[i].start;
buf.length = buffers[i].length;------------------->>>>>>if (-1 == xioctl(fd, VIDIOC_QBUF, &buf))
errno_exit("VIDIOC_QBUF");
}
type = V4L2_BUF_TYPE_VIDEO_CAPTURE;
if (-1 == xioctl(fd, VIDIOC_STREAMON, &type))
errno_exit("VIDIOC_STREAMON");
break;
}Можете пояснить?
> Спасибо за ответ.
> Делал по примеру отсюда: https://linuxtv.org/downloads/v4l-dvb-apis/capture-example.html...
> Можете пояснить?
- давно несколько в другом направлении копаюсь
- с cuda вообще не пересекался - есть возможность работы с памятью выделенной через стандартный malloc или надо строго через cudaMallocHost( pinned, size ) работать?
- как вариант(?), проверка на ошибки пропущена, кусок кода оттуда жеif (-1 == xioctl(fd, VIDIOC_DQBUF, &buf)) {
switch (errno) {
case EAGAIN:
return 0;case EIO:
/* Could ignore EIO, see spec. *//* fall through */
default:
errno_exit("VIDIOC_DQBUF");
}
}
Разумеется буферы выделены корректно, проверяю.
До Вашего куска не доходит.
Зависает в start_capturing( в режиме IO_METHOD_USERPTR)Да, надо в cuda иначе все совсем медленно.
Сорри поторопился.
Сразу проверить код ошибки в start_capturing не пробовал.
Попробую отпишусь.
> Сорри поторопился.
> Сразу проверить код ошибки в start_capturing не пробовал.
> Попробую отпишусь.пока идет локальная пересборка,
ради интереса глянул в сети инфу по cuda,
http://ecee.colorado.edu/~siewerts/extra/code/example_code_a.../немного статистики
14 страница в CUDA_Getting_Started_Guide_For_Linux.pdfGTX 670, memory pinned:
host2device: 3 G/s
device2host: 2 G/s+
имхается, надо глянуть примеры:
.../cuda_work/samples/0_Simple/simpleStreams/
.../cuda_work/samples/0_Simple/simpleZeroCopy/
P.S.:а ус-во видео захвата на какой шине сидит:
USB(2,3), PCI(32,64), etc ?
Прибор Jetson TK1, немного солгал про Tegra 3, наверное Tegra 4.
В вики пишут, что 64бит шина. Причем RAM у GPU и CPU общая, отсюда еще больше вопросов к memcpy. 200мс/26Мб это 130Мб/сек. SSD быстрее пишет :)
На выхах буду анализировать ссылки и ответы присланные вами.
Как вариант, запланировал эксперимент про мультипоточность, т.к. в момент memcpy занято одно ядро из 4-х.
Отпишусь по результатам.
> Прибор Jetson TK1, немного солгал про Tegra 3, наверное Tegra 4.
> В вики пишут, что 64бит шина. Причем RAM у GPU и CPU
> общая, отсюда еще больше вопросов к memcpy. 200мс/26Мб это 130Мб/сек. SSD
> быстрее пишет :)...
думается, сырой захват идет через USB 2
+
как вариант, видео "пожатое" идет через тотже USB..
имхается, поэтому такая задержка...
Камера через MIPI CS2 4lane. Тут все нормуль. Данные в yuv422.
> Камера через MIPI CS2 4lane. Тут все нормуль. Данные в yuv422.bandwidth ? (dma ?)
yuv422 - это raw data
Реально не знаю.
По вики много.
Пробовал memcpy из простого буфера в простой:uchar * pIn{ new uchar[ size ]};
uchar * pOut{ new uchar[ size ]};
memcpy( pOut, pIn, size );
delete [] pIn;
delete [] pOut;Тоже около 200мс.
Пока получается, что на Tegra TK1 4224х3156х2 байт копируются с такой скоростью.
Но это учитывая, то, что там драйвер камеры и подсистема V4L2 делает.
Буду разбираться.Сделал вариант с OpenMP, реально на 4 поделилось время:
static void memcpy_openmp( uchar * pOut, const uchar * pIn, const int size ) {
//omp_set_dynamic( 0 );
omp_set_num_threads( 4 );
//int blockSize{ 1666368 };
int blockSize{ size / 4 };
int th_id, nthreads;
#pragma omp parallel for
for( int i = 0; i < size / blockSize; i++ ){
int offset{ i * blockSize };
memcpy( pOut + offset, pIn + offset, blockSize );
th_id = omp_get_thread_num();
nthreads = omp_get_num_threads();
qDebug() << th_id << nthreads;
}}
Где-то 70-80мс.
Вопрос открыт.
По V4L2:https://devtalk.nvidia.com/default/topic/894783/jetson-tk1/v...-/
IO_METHOD_USERPTR не поддерживается. Только MMAP.
> По V4L2:https://devtalk.nvidia.com/default/topic/894783/jetson-tk1/v...-/
> IO_METHOD_USERPTR не поддерживается. Только MMAP.остаются: v4l2_memory_mmap & v4l2_memory_dmabuf
может что-то из этого зафурычит..
Спасибо. Все ваши советы обязательно проверю. Не было времени пока.
v4l2_memory_dmabuf не поддерживается.Только MMAP.
Пока сделал параллельно. Укладываюсь в 70мс. Но позже надо будет в mp4 жать, надеюсь CUDA на себя все возьмет.
Сделал тест:- memcpy from userspace pointer to userspace pointer is about 33ms per 26mb
- memcpy from userspace pointer to pinned pointer is about 33ms per 26mb
- memcpy from mmap v4l pointer to pinned is about 200ms per 26mbMmap медленный. Интересно почему..
Тут у человека такая же проблема: https://devtalk.nvidia.com/default/topic/948258/jetson-tk1/p.../
Решилось с помощью V4L2_MEMORY_USERPTR.В найденном варианте v4l2 нашел почти похожий пример(http://www.friendlyarm.net/forum/topic/1006), как в стандартном примере, но пользовательский буфер выделяется не malloc, а memalign из malloc.h. Иначе Invalid argument error 22.
Настраиваемся на IO_METHOD_USERPTR и в init_userp:
unsigned int page_size;
page_size = getpagesize ();
buffer_size = (buffer_size + page_size - 1) & ~(page_size - 1);..................
..................for (n_buffers = 0; n_buffers < 4; ++n_buffers) {
buffers[n_buffers].length = buffer_size;
buffers[n_buffers].start = memalign (/* boundary */page_size,buffer_size);if (!buffers[n_buffers].start) {
fprintf (stderr, "Out of memory\n");
exit (EXIT_FAILURE);
}
}
И далее работаем с буфером buf.m.userptr размера buf.length.Что в отличии от mmap буфера, позволяет быстро копировать данные кадра в pinned memory и далее в cuda.
Думаю что вопрос закрыт.
Всем спасибо, особенно пользователю fail :-)
С уважением Виктор.
> Решилось с помощью V4L2_MEMORY_USERPTR.
> В найденном варианте v4l2 нашел почти похожий пример(http://www.friendlyarm.net/forum/topic/1006),
> как в стандартном примере, но пользовательский буфер выделяется не malloc, а
> memalign из malloc.h. Иначе Invalid argument error 22.мда,
насчет memory alignment - совсем из головы вылетело - хотя:- специализированные {C,G}PU
- прочая эмбедовкак этому фактору бывают очень чуствительны..
...