Хочу выбрать IDE для CUDA, работаю под Kubuntu, больше всего нравится Eclipse CDT, однако два найденных в интернете руководства не смог осилить, делаю всё по инструкции, но не получается результат как у авторов, к тому же они привязываются к CUDA SDK, к его путям и make-файлу common.mk, где всё на нем завязано, а мне бы хотелось использовать свои.
С NetBeans тоже не получилось, там еще меньше инфы нашел. Есть реализация FindCUDA для CMake, но мне бы хотелось нормальную среду с эдитором, спейсингом и дебагом, а как в cmake дебажить, не знаю, консольный дебаг-не вариант.
Подскажите, кто на чем работает, есть ли работающие инструкции?
Собственно, задача - переписать существующее ПО цифровой обработки изображений под CUDA. Ранее использовалась библиотека GENIAL, удобная для работы с матрицами, можно было работать местами почти как в MATLAB, операции +,-,*,/ работали с матрицами как и с числами.
Однако посмотрев на классическую CUDA, понял, что там всё не так устроено, хотелось бы более удобный вариант, чтоб акцент был на функциональной части. CUBLAS - не подходит, мощные функции, а простые трудно реализовать. Есть "обертки" над CUBLAS. Я обратил внимание на две:
Thrust - обертка , аналогичная по интерфейсу STL, местами удобная, но немного странная как для меня
VSIPL - мощная библиотека, судя по описанию, использует параллельные вычисления, в том числе и CUDA, но явно не указано, в какой степени оно зависит от CUDA, функций великое множество, ближе всего по реализации к GENIAL.
Если кто-то ими интересовался или использовал, или знает другую библиотеку, подскажите, пожалуйста.
Или может имеет смысл перейти на OpenCL, правда не знаю, насколько он сейчас самостоятелен. Критериями является быстродействие и удобство работы с матрицами, наличие простых и сложных операций и функций.
IDE для CUDA
By blackswan - October 20th, 2009
Tagged:
Comments
вроде бы собрал, на основе template проекта из SDK, всё получилось, однако мне кроме CUDA надо подключить и обычный код для загрузки PNG файлов, в частности хедер и статик-либрари, однако пока что сыпяться ошибки. Вот мой makefile
EXECUTABLE := ../../../src/projectc/build/ptest
CUFILES := projectc.cu
CU_DEPS := matutils.cu
CCFILES := image.cpp
include ../../common/common.mk
CXXFLAGS += -I/usr/include -I/usr/local/include -I. -I..
CXXFLAGS += -L/usr/lib -L/usr/local/lib -L/usr/local/cuda/lib/ -O3
LDFLAGS += -lpng
где project.cu - основной файл с main функцией, matutils.cu - файл с функциями CUDA, image.cpp - файл с обычными функциями по загрузке PNG изображения. Только пока то, что в image.cpp ошибки, так как он не видит хедеры stdio.h png.h и т.д., которые я вставил в projectc.cu , а именно:
#include "stdlib.h"
#include "stdio.h"
#include "string.h"
#include "math.h"
#include "time.h"
#include "png.h"
using namespace std
#include
#include
#include
#include
С чем это связано? Файлы с расширением .cu полностью отдаются на компиляцию в CUDA или там можно писать код обычного С++?
так вроде разобрался. КУДе кудово, если файл с .cu, то хедеры из этого файла не транслируются в файлы cpp и они их не видят, я прав? Теперь не могу отладить функцию матричного умножения (не работает, но ошибок не выдает), как и другие, написанные для видеокарты. Кто какими методами дебага пользуется?
Во-первых: при компелции в режиме эмуляции(для компиляции с использованием common.mk параметр emu=1) можно использовать например printf даже в __device__ функции. Можно использовать расширение гнушного отладчика cuda-gdb.
компилирую (make emu=1 dbg=1), однако в функциях __global__ printf не работает, ничего не выдает, пример
__global__ void __sadd(float * a, float b, float * c, long N, long M)
{
long i = blockIdx.x * blockDim.x + threadIdx.x;
long j = blockIdx.y * blockDim.y + threadIdx.y;
long index = i + j * N;
printf("%d %d\n",i,j);
if( i < N && j < M)
c[index] = a[index] + b;
__syncthreads();
}
а отладчик cuda-gdb можно прикрепить к Eclipse? Или как там ставить брейкпоинт? Из сухого мануала, что я нашел, мало понятного.
Да еще __syncthreads(); в нутри ветвления НЕ НАДО ставить только если ты уверен что все потоки в блоке пройдут по одномой ветке. Иначе у тебя пожет программа уйти в себя )). У барьера ожидаться будут потоки, которые в эту ветку никогда не зайдут!
В Eclipse у меня не получилось интегрировать cuda-gdb.
Для использования cuda-gdb надо иметь представление о работе gdb. Вообще про него(gdb) вот тут очень хорошо написано http://www.linux.org.ru/books/GNU/gdb/gdb_toc.html. А в Nvidia_CUDA_GDB_User_Manual_2.3 есть пошаговый пример(но сначала GNU/gdb) - думаю его мысла нет тут переписывать :-)
Хотя проблема вроде бы не требует исползования cuda-gdb.
Насчет отладчика не знаю, но функция printf, насколько мне известно, в gpu коде и не должна работать. Все что в кернеле компилируется под GPU.
В режиме эмуляции - должна.
да, заработал. Однако всё равно не могу вычислить , почему не работает матричное умножение. Использую код с сайт NVidia, Моя задача - умножить матрицу изображения, представленную как 3 канала Х Высота*Ширина, каждый канал на R,G и B соответственно, вытянуты они в одну строку, на другую матрицу константных чисел, размером всего лишь 3Х3 для преобразования в YIQ формат. По правилу необходимо умножить так T * ImageRGB = ImageYIQ, где T[3X3] и ImageRGB[3XHeight*Width], чтобы размер ImageYIQ остался таким же, как ImageRGB. однако функция умножения таким образом не запускается (пишу в ней printf для проверки входа). А когда проверяю умножения на других размерах, например [3*HeightXWidth], то умножение работает, однако мне надо именно [3XHeight*Width]. Исходя из кода функции я понял, что умножение разбивается на блоки-подматрицы размером BLOCK_SIZE X BLOCKSIZE, чтоб использовать в каждом блоке shared memory. Я чувствую подвох в этом, поэтому выставил BLOCK_SIZE = 1 вместо 16, чтоб не получилось 3/BLOCK_SIZE = 0. Однако это не помогло, а разобраться в расчете индексов входных и выходной матриц и их подматриц, принятом в функции пока что не получается. У кого-нибудь есть идеи или другой пример матричного умножения на CUDA?
Пока что хочу разобраться без оптимизации. Оказалось, что функция работает, однако только для маленьких изображений, т.е. проверял на нескольких, в итоге на изображении в 400Х400 и выше уже не работает, совсем не запускается, а на тех, что меньше, запускается и туда заходит. Как это связано с размером, не могу понять. Вот еще раз код функции
BLOCK_SIZE равен 1, изначально был равен 16, но я изменил, чтобы предотвратить 3/BLOCK_SIZE = 0.
Я настоятельно рекомендую прочитать про аппаратные особенности видокарты. Хотя бы внимательно прочитать слайды с МГУ-шной презентации.
да, действительно, проблема в этом. Пока что представил каждый канал в виде отдельной матрицы [Height X Width] и всё заработало. Вот теперь не знаю, как организовать вычисление среднего арифметического с использование CUDA, там аккумулирующая сумма - одна и та же ячейка в памяти, каждый поток одновременно или через гонки будет пытаться получить к ней доступ, и что они там насуммируют - неизвестно, результат должен быть ошибочным. Задача мне кажется исключительно последовательной. Как её решать в GPU?
Посмотрите как сделаны гистограммы (в примерах CUDA там даже дока отдельная на это место есть), это если хочется совсем выпендриться.
Но вообще, все несложно. Как-то так:
- делаем не очень много блоков (скажем 128)
- в каждом делаем не очень много threads
- в хвосте thread обычной редукцией складываем среднее для блока и записываем в глобальную память (уникальное место для каждого блока)
На хосте усредняем оставшиеся 128 чисел.
Под "обычной редукцией" вы понимаете каскадную схему суммирования?
есть еще одно, у меня иногда проявляются артефакты на экране в виде "снега", видимо портится память в видеокарте. Как вообще она там выделяется при выполнении программы? возможно ли , что пересечение памяти под другие приложения (в том числе отрисовки оболочки) и памяти программы?
я открыл код reduction из toolkit, однако такой код мне не очень понятен, я написал свой, видимо это тоже каскадное суммирование, он работает, однако мне кажется, что выдает неверные значения, так как есть реализация обычная для тех же данных (того же изображения), и эти ср. арифметические отличаются от тех. Я разбил задачу на две стадии, сделав предварительный массив сумм.
всё проверил, похоже этот код среднего арифметического правильный, но не совсем. Работает он корректно только в эмуляции, когда же пытаюсь исполнить на самой видеокарте, он выдаёт неадекватные числа, или вообще создаёт артефакты на экране. Тут как будто две проблемы, рассинхронизация потоков и что-то с памятью видеокарты. Но вот конкретно не могу понять, где же это возникает.
Боюсь огорчить, но обычно людям не свойственно читать чужой код с экрана и пытаться выполнить его "в уме".
:-D
я прошу совета, а не компилировать и исполнять мой код у себя. Возможно у кого-то были подобные ситуации и ошибки. Я нашел презентацию Харриса по редукции, взял оттуда метод под номером 4 (Reduction #4: First Add During Load), однако он вызывает segmentation fault. Там два параметра - входная и выходная матрица. Пробовал туда передавать одну и ту же и две разных матрицы, всё равно вываливается.
Я бы попробовал __meanStage1 и __meanStage2 выгрузить sum в режиме эмуляции и без него и сравнить их, тем самым проверить правельно ли работает __meanStage1. Так хоть можно будет чуть локализовать ошибку
Когда изображение не превышает (примерно) 400Х400, обе работают нормально в режиме эмуляции и без. А когда большие изображения, то неправильно работает __meanStage1, при эмуляции даёт неверные данные, при обычном режиме бывает неверные данные, а бывает экран покрывается "артефактами". Я пробовал разные "высота х ширина", но закономерность вижу пока только в увеличении размеров.
Если работает в эмуляторе (особенно если эмулятор исполняется на одном ядре/процессоре), но не работает на устройстве - скорее всего у вас проблемы с синхронизацией, разные треды пишут в одну shared mem или разные треды/блоки - в одно место глобальной памяти.
За минуту я проблемы не увидел, но симптомы - такие.
Надо бы пренести то что не касается IDE в другую тему.
Ну да.
Посмотрите пример из CUDA Toolkit, так и называется reduction, там в doc/ лежит презентация Харриса "как мы ускорили редукцию в 30 раз"
Одна из ваших проблем связана с тем, что размерность грида равна iH*iW т.е. размеру картинки, для картинок больше чем 256x256 вы превышаете максимальную размерность в одном измерении (65536 или около того)
Ставьте тег code, а то больше-меньше пропадают.
А чему равен BLOCK_SIZE?
> однако функция умножения таким образом не запускается
Какая ошибка выводится?
>Я чувствую подвох в этом, поэтому выставил BLOCK_SIZE = 1 вместо 16
Никто же не запрещает сделать 2 переменные BLOCK_SIZE_X = 3 BLOCK_SIZE_Y = 16(или другие значения).
А в кернале поспаить проверку что бы лишние значения не считались как то так например:
Где x_end и y_end размер ImageRGB
Примичания: 1) Конечно не стоит забывать что потоки не попавшие в if будут простивать 2) Блок потоков размером 3Х16 - плохой вариант. Если уж BLOCK_SIZE_Х = 3 - жёстко то BLOCK_SIZE_Y я бы сделал равным 85(3*85 = 255 потоков). Но это уже дело оптимизации
3x3 можно и руками расписать
На мой взгляд использование common.mk в своих make-файлах достаточно удобно под Linux. Eclipse использую только для раскраски текста). Но то что GNU-ый отладчик не самый удобный - соглашусь! Хотя это хотя бы что-то.
меня и так тоже устроит, только не знаю, как работать с common.mk. Он используется для make? Я не имел опыта работы с make напрямую. Все привык делать с визуальными IDE. Может расскажете, как её использовать и настроить для CUDA? А с gdb тоже работаете в консоли? Ведь он не позволяет отлаживать код функций CUDA или позволяет? У куда ведь был свой отладчик.
да, пока 0 ответов. Все же на чем-то пишут. Я взял findCUDA-проект, правда ничего не понимаю в CMAke и по одной инструкции сделал связку CMake и Eclipse, где в Eclipse создан проект с внешним makefile - ом, который генерируется как раз с помощью CMake. Причем CMake был добавлен как external tool приложение, которое я сконфигурировал и запускаю, когда нужно переделать makefile, после того как я редактирую файл CMakeList.txt. Первое приложение я написал на GPU VSIPL, при этом всё прекрасно компилилось и билдилось, даже получил прирост в 5 раз по сравнению с обычным исполнением на CPU. По идее он должен был использовать CUDA, так заявляют разработчики, однако я решил перейти на чистую CUDA, так как мне не понравился такой слабый прирост GPU VSIPL, и тут столкнулся с проблемами. У меня выпадают вот такие ошибки при билде
/matutils.cu:112: ошибка: нет декларации blockIdx в этой области видимости
/matutils.cu:112: ошибка: нет декларации blockDim в этой области видимости
/matutils.cu:112: ошибка: нет декларации threadIdx в этой области видимости
/matutils.cu:117: ошибка: нет декларации __syncthreads в этой области видимости
и т.д. по поводу самых основных объектов CUDA, я подключил два хедера,
#include "/usr/local/cuda/include/cuda.h"
#include "/usr/local/cuda/include/cuda_runtime.h"
их должно быть достаточно, остальные должны были подключится из них, однако ни Eclipse, ни CMake похоже не видят эти хедеры, или же мне еще подсказали на буржуйском форуме, что я всё еще не использую nvcc для компиляции, но я не знаю, где он настраивается, в Eclipse в свойствах проекта ничего нет, так как он предполагает использование стороннего Makefile и там должен быть nvcc, в CMakeList.txt тоже не вижу упоминания про nvcc, не знаю как его туда (или куда) впихнуть. Подскажите, кто на чём работает? Вообще у меня Ubuntu 9.04 и на другую ОС перейти не могу.