В същото спецификатори __host__ __device__ и може да се използва заедно (което означава, че съответната функция може да се извърши на графичния процесор, и процесора - на съответния код ще се генерира автоматично от компилатора за двете платформи). Проектантите __global__ __host__ и не могат да се използват заедно.
__global__ спецификатор обозначава ядрото и съответната функция трябва да връща стойност от тип нищожно.
__global__ невалидни MYKERNEL (поплавък * а, поплавък * б, поплавък * в)
индекс Int = threadIdx.x;
На функциите, изпълнявани в графичния процесор (__device__ и __global__) Прилагат се следните ограничения:
За да укажете местоположението в паметта на GPU променливи, следните спецификатори - __device__, __constant__ и __shared__. Използването им налага редица ограничения:
В езика на следните специални променливи се добавят:
- gridDim - размер grid'a (тип dim3)
- размер блок (тип dim3) - blockDim
- blockIdx - текущия блок индекс grid'e (е тип uint3)
- threadIdx - индексът на текущата нишка в блока (вид uint3)
- warpSize - размер warp'a (вид на междинно съединение)
Езикът добавя 1/2/3/4-двумерен вектор на основните видове - ЗНАК1, ЗНАК2, char3, char4, uchar1, uchar2, uchar3, uchar4, short1, short2, short3, short4, ushort1, ushort2, ushort3, ushort4, INT1 , int2, int3, int4, uint1, uint2, uint3, uint4, long1, long2, long3, long4, ulong1, ulong2, ulong3, ulong4, float1, float2, float3, float2 и double2.
Свързване на компонентите на вектора са имената - X, Y, Z и W. За да създадете векторите от стойности на даден тип е тип дизайн make_ на
int2 а = make_int2 (1, 7);
float3 ф = make_float3 (1, 2, 3.4f);
Моля, имайте предвид, че за тези видове (за разлика от засенчване език GLSL / С $ / HLSL) не се поддържат векторни според компонентите, операции, т.е. не може просто да се добавят два вектора с използване на оператора "+" - очевидно е необходимо да се направи за всеки компонент.
Също така служи за задаване на DIM3 тип величина, въз основа на uint3 тип, но има един нормален конструктор инициализира всички компоненти не са определени единици.
директива ядрото повикване
За да стартирате ядрото на GPU използва следната конструкция:
kernelName <<
Също така в езикови добавена __syncthreads функцията C синхронизира всички нишки на блока. извън контрол й ще бъде върнат само, когато всички нишки на даден блок ще доведе до тази функция. Т.е. когато всички кодът, който върви пред настоящата покана, вече са направени (и, следователно, можем спокойно да разчитате на резултатите). Тази функция е много удобна за работа на организацията bezkonfliktnoy с споделена памет.
CUDA поддържа всички математически функции на стандартна библиотека, но от гледна точка на ефективността на по-доброто използване на техните поплавък аналози (не се удвои) - например sinf. В допълнение, CUDA осигурява допълнителен набор от математически функции (__sinf, __powf и т.н.) осигурява по-ниска точност, но значително по-висока производителност в сравнение с sinf, powf т.н.
Основи CUDA домакин API
CUDA API за CPU (домакин) се появява в две форми - ниско ниво CUDA drievr API една и CUDA изпълнение API (осъществят с помощта на водача програмен интерфейс на CUDA). В молбата си, можете да използвате само един от тях, а след това ние ще разгледаме по време на работа CUDA API, толкова по-лесен и удобен.
Всички функции CUDA шофьор API куб.м започва с префикс и всички функции на CUDA изпълнение API започват с CUDA префикс. Всеки един от тези API предоставя основен набор от основни функции, като например чрез всички налични устройства (GPU), работа с контексти и конци работа с паметта на графичния процесор, взаимодействието с OpenGL и D3D (само с подкрепата на 9-ия версия на DirectX).
CUDA изпълнение API не изисква изрично инициализация - това се случва автоматично, когато първата покана за някой от неговите функции. Важен аспект от работата с CUDA е, че много от функциите на API са асинхронни, т.е. контрол се връща преди самото завършване на необходимата операция.
Сред асинхронни операции включват:
- ядро старта
- функции за копиране на паметта, чиито имена завършват на асинхронен
- Memory функция копие устройство <-> приспособление
- функция памет инициализация.
CUDA поддържа синхронизация чрез потоци (струи) - всеки поток определя последователността на операциите, извършвани в строг ред. Когато този ред на операциите между различните потоци не е строго определена, и може да варира.
Всяка функция CUDA API (с изключение започва ядрото) връща тип cudaError_t. При успех, функцията връща cudaSuccess, в противен случай връща код за грешка.
Вземете описание на грешката във формата на неговия ред код може да бъде използване funktsiicudaGetErrorString:
знак * cudaGetErrorString (cudaError_t код);
Можете също така да получите последната код за грешка, използвайки funktsiicudaGetLastError:
Моля, обърнете внимание, че поради асинхронно изпълнение на много разговори, за да получите кода за грешка е по-добре да се използва функция cudaThreadSynchronize, която е в очакване на приключване на-всички GPU предавани искания и връща грешка, ако едно от тези искания е довело до грешката.
Работа с памет в CUDA
Най-лесният начин да се разпределят и свободна памет (ние говорим само за паметта на графичния процесор, като само линейна памет, друг тип - CUDA масиви ще бъдат обсъдени в следващата статия) е да се използва cudaMalloc функции, cudeMallocPitch и cudaFree.
поплавък * devPtr; // указател към паметта на устройството
// разпределят линеен памет за 256 поплавъци
cudaMalloc ((невалидни **) devPtr, 256 * sizeof (флоат));
За определянето на способността CUDA използва понятието Изчислява способност, изразена от двойка номера - MAJOR.MINOR. Първото число се отнася до глобалната архитектурна версия, а вторият - малка промяна. Тъй GeForce на графичния процесор 8800 Ultra / GTX / GTS са равни Изчислява Възможност 1.0, графичния процесор GeForce 8800 GT / GS и GeForce 9600 GT са равни Изчислява Възможност 1.1, графичния процесор GeForce GTX 260 и GeForce GTX 280 равен Изчислява Възможност 1.3.
Compute Capability 1.1 поддържа атомни операции на 32-битови думи в глобалната памет, Compute Capability 1.2 поддържа атомни операции в споделена памет и атомни операции на 64-битови думи в глобалната памет, Compute Capability 1.3 поддържа операции на номерата на двойно тип.
По-долу е сорс кода на една проста програма, която показва всички налични GPU и техните основни характеристики.
Int основна (междинно argc, знак * argv [])
ФОРМАТ ( "намерени устройства \ н.", deviceCount);
за (междинно устройство = 0; устройството cudaGetDeviceProperties ( devProp устройство); ФОРМАТ ( "Устройство \ н.", устройство); ФОРМАТ ( "Compute способност .... \ Н", devProp.major, devProp.minor); ФОРМАТ ( "Name% и \ н.", devProp.name); ФОРМАТ ( "Общ Global \ Memory н ..", devProp.totalGlobalMem); ФОРМАТ ( "Обща памет на блок: \ н.", devProp.sharedMemPerBlock); ФОРМАТ ( "апарати блок \ п на ..", devProp.regsPerBlock); ФОРМАТ ( "Warp размер \ Н ..", devProp.warpSize); ФОРМАТ ( "Максимален нишки блок \ п на ..", devProp.maxThreadsPerBlock); ФОРМАТ ( "Общ постоянна памет \ Н ..", devProp.totalConstMem); Помислете за някои прости примери за използване на CUDA да демонстрира основните техники на работа с нея. Най-простият пример е просто увеличаване на всеки елемент едномерен масив от един - incr.cu. програма __global__ невалидни incKernel (поплавък * данни) Най-лесният начин е разположена ядро - един с резба, съответстваща резба блокове и решетка двумерен. сърцевина (incKernel функция) вход получава само указател към масив от данни в световен памет. Целева ядро - за threadIdx blockIdx и да определи кой елемент съответства на резбата и да се увеличи неговото име. Тъй като и двете блоковете и едномерен мрежата, броят на прежда се определя като броя на блок, умножена по броя на влакна в блока, плюс броя резба в блок, т.е. blockIdx.x * blockDim.x + threadIdx.x. Основната функция е малко по-сложно - той трябва да се подготви набор от данни в паметта на процесора, а след това трябва да използвате cudaMalloc памет за копие на масива с данни в глобалната памет (DRAM GPU). На следващо място, данните се копират функция cudaMemcpy от CPU памет на глобалната GPU паметта на. След приключване на копирането на данни от глобалната памет може да започне за данни, които ядрото след призива му да копирате обратно изчисления резултат на GPU глобалната памет в CPU памет. Размножаване на две матрици - простият подход Следващият пример е по-сложно (и спешно) - ние ще разгледаме използването на CUDA за умножаване на две квадратни матрици с размер N * Н. Да предположим, че имаме две квадратни матрици А и В с размер N * N (приемем, че N е кратно на 16). Най-простото изпълнение използва един конец за всеки елемент на получената матрица С, при което резбата извлича всички необходими елементи на глобалната памет, и извършва необходимите изчисления. Елемент CI, J продукта от две матрици А и В се изчислява по следния псевдо код фрагмент: По този начин за изчисляване на един елемент на продукта за матрица трябва да изпълни 2 * N аритметика и 2 * N отчитания от глобалната памет. Ясно е, че в този случай, основният ограничаващ фактор е скоростта на достъп до глобалната памет, което е много ниско. Как личната нишки групирани в блокове не е важно и не оказва съществено влияние върху резултатите от дейността, която в този случай е много ниска. По-долу е списък на съответната програма. Размножаване на две матрици с помощта на обща памет Може значително да се подобри ефективността на нашите програми чрез използване на споделена памет. За тази цел, се разделят на получената матрица в подматрици 16 * 16, изчисляване на всяка от тези подматрица ще бъдат ангажирани в един блок. Моля, имайте предвид, че при изчисляването на такъв подматрица трябва само малко "банда" на матрици А и Б. За съжаление изцяло репликира тези "ивици" в споделената-паметта е практически невъзможно поради относително малкия размер на споделена памет. Поради това е възможно да се направи по друг начин - разделихме "лентата" на матрицата 16 * 16 и изчисляването на под-матрица на продукта матрицата ще се извършва в N / 16 стъпки. За тази цел, ние се отбележи, че при изчисляването на ДИ елемент, й може да бъде пренаписана, както следва, като се използват преградни решетки в квадратен подматрица за стъпка в 0..N / 16: в [Ь] [й] + = на [Ь] [К + стъпка * 16] * б [К + стъпка * 16] [й] Имайте предвид, че за всяка стойност на стойността на стъпка на матрици А и В е взета от две подматрици с размер 16 * 16. В действителност, групата с ориз. 4 просто разделен на квадратен подматрица и всяка стойност съответства на една стъпка на такива подматрица А и един подматрица Б. На всяка стъпка ние се зареди в споделена памет на 16 * 16 подматрица А и един 16 * 16 подматрица Б. След това, ние се изчисли необходимото количество от тях за елементи на продукта и след това заредете следващите 16 * 16 подматрици и т.н. На всеки етап, една нишка товар точно един елемент на всяка от матрици А и В и изчислява сумата на съответните им членове. Когато всички изчисления се извършват запис елемент в крайния матрицата. Моля, имайте предвид, че след изтеглянето на елементите на А и Б е необходимо да се синхронизира теми, като се обадите __synchronize до момента на началото на селища от всички необходими елементи (сваляне останалите главни теми) би вече са заредени. По същия начин, в края на лечението са били добавени подматрици и се нуждаят от синхронизация, преди да изтеглите те също (за да се уверите, че сегашната 16 * 16 подматрица вече не е необходимо и възможно ново зареждане е). Следва съответния изходен код.
Всичко за Football Manager
проекти Новини
Какво момента работим по проекти
Проектни планове, какво да правя и просто планове
Развитие, че след като е направил, или правя, както и мисли и бележки.
уеб дизайн
Програми за развитие или programmulek, функции или блокове.
Различни бележки, идеи за проекти.
Всичко, което не спада в предишната колона
Новини от света на факти.
Преглед на програми, представляващи интерес за моя ум
Преглед на желязо в света
Различни статии за програмиране, разработване на приложения.облак тагове
KarpOlya - нотки на учителя по биология и химия
Онлайн футболен мениджър
Свързани статии