>Это только одна часть.
Я бы сказал вводная.
>К ней есть продолжение, возможно будут следующие статьи, по крайней мере работу с текстурной памятью
Это нужно. Даже весьма.
>и OpenGL лучше все-таки вынести в отдельную статью, ане добавлять в эту.
OpenGL есть куча статей, но интересно именно взаимодействие с cuda.
>А что подразумевается под п2 ?
В основном совместная работа c++, cuda и какого-нибудь графического апи.
Я прослушал аудиолекции читаные на мгу и могу выразить общие впечатления.
Первые три лекции для меня (наверное не только для меня) оказались самыми полезными, там было детально разжевана работа с памятью.
Последующие же лекции, хоть они и затрагивали вопросы мне интересные, но к сожалению без видеокартинки следить за ними очень трудно (когда лектор говорит: вот эта строка - тут я пас, какая именно строка, могу только догадываться). Кроме того, последние лекции это не о куде, а скорее рассказ авторов, о своей работе на куде. Что совсем не одно и то же.
Кстати с замечанием, что работать со встроеными типами cuda: float2, float3 и т.д. ручками, покомпонентно, я никак согласится не могу. Cuda поддерживает перегрузку операторов, в том числе и на девайсе.
Вот например моя векторная билеотека для работы со стандартными типами Cuda.
Код:#ifndef _VECTOR_CUH_
#define _VECTOR_CUH_
#include <vector_types.h> //ñòàíäàðòíûå âåêòîðíûå òèïû CUDA
template <typename T>
__device__ T swap(T &a, T &b)
{
T temp;
temp = a;
a = b;
b = temp;
}
//Øàáëîíûå ïðåîáðàçîâàíèÿ ìàññèâà float â ñòàíäàðòíûå âåêòîðíûå òèïû CUDA
template <typename T>
__device__ T makeFromAr(const float *v)
{
T res;
return res;
}
template <>
__device__ float4 makeFromAr<float4>(const float *v)
{return make_float4(*v, *(v+1), *(v+2), *(v+2));}
template <>
__device__ float3 makeFromAr<float3>(const float *v)
{return make_float3(*v, *(v+1), *(v+2));}
template <>
__device__ float2 makeFromAr<float2>(const float *v)
{return make_float2(*v, *(v+1));}
//Îïðåäåëåíèå àðèôìåòè÷åñêèõ îïåðàòîðîâ:
//Âåêòîðíîå ïðîèçâåäåíèå:
__device__ float3 operator ^(const float3 &a, const float3 &b)
{return make_float3(a.y*b.z-a.z*b.y, a.z*b.x-a.x*b.z, a.x*b.y-a.y*b.x);}
//Âåêòîðíîå ïðîèçâåäåíèå (äëÿ 2D ïîíèìàåòñÿ êàê äëÿ ïëîñêîïàðàëåëüíîé ñèñòåìû):
__device__ float operator ^(const float2 &a, const float2 &b)
{return a.x*b.y-a.y*b.x;}
//âåêòîðíîå óìíîæàíèå âåêòîðà arg ëåæàùåãî íà ïëîñêîñòè XY íà îðò k (îñè Z)
__device__ float2 CrossProductOn_k(float2 arg)
{
return make_float2(arg.y, -arg.x);
};
//Íîðìà:
__device__ float Norm(const float4 &a)
{return (fabs(a.x)+fabs(a.y)+fabs(a.z)+fabs(a.w))*.25f;}
__device__ float Norm(const float3 &a)
{return (fabs(a.x)+fabs(a.y)+fabs(a.z))*.33333333333f;}
__device__ float Norm(const float2 &a)
{return (fabs(a.x)+fabs(a.y))*.5f;}
//Äëèíà:
__device__ float Length(const float4 &a)
{return sqrtf(a.x*a.x+a.y*a.y+a.z*a.z+a.w*a.w);}
__device__ float Length(const float3 &a)
{return sqrtf(a.x*a.x+a.y*a.y+a.z*a.z);}
__device__ float Length(const float2 &a)
{return sqrtf(a.x*a.x+a.y*a.y);}
//Êâàäðàò äëèíû:
__device__ float Length2(const float4 &a)
{return a.x*a.x+a.y*a.y+a.z*a.z+a.w*a.w;}
__device__ float Length2(const float3 &a)
{return a.x*a.x+a.y*a.y+a.z*a.z;}
__device__ float Length2(const float2 &a)
{return a.x*a.x+a.y*a.y;}
//Ñóììà:
__device__ float4 operator+(const float4 &a, const float4 &b)
{return make_float4(a.x+b.x, a.y+b.y, a.z+b.z, a.w+b.w);}
__device__ float3 operator+(const float3 &a, const float3 &b)
{return make_float3(a.x+b.x, a.y+b.y, a.z+b.z);}
__device__ float2 operator+(const float2 &a, const float2 &b)
{return make_float2(a.x+b.x, a.y+b.y);}
//Ðàçíîñòü:
__device__ float4 operator-(const float4 &a, const float4 &b)
{return make_float4(a.x-b.x, a.y-b.y, a.z-b.z, a.w-b.w);}
__device__ float3 operator-(const float3 &a, const float3 &b)
{return make_float3(a.x-b.x, a.y-b.y, a.z-b.z);}
__device__ float2 operator-(const float2 &a, const float2 &b)
{return make_float2(a.x-b.x, a.y-b.y);}
//Ñêàëÿðíîå ïðîèçâåäåíèå:
__device__ float operator|(const float4 &a, const float4 &b)
{return a.x*b.x+a.y*b.y+a.z*b.z+a.w*b.w;}
__device__ float operator|(const float3 &a, const float3 &b)
{return a.x*b.x+a.y*b.y+a.z*b.z;}
__device__ float operator|(const float2 &a, const float2 &b)
{return a.x*b.x+a.y*b.y;}
//Ïî÷ëåííîå ïðîèçâåäåíèå:
__device__ float4 operator%(const float4 &a, const float4 &b)
{return make_float4(a.x*b.x, a.y*b.y, a.z*b.z, a.w*b.w);}
__device__ float3 operator%(const float3 &a, const float3 &b)
{return make_float3(a.x*b.x, a.y*b.y, a.z*b.z);}
__device__ float2 operator%(const float2 &a, const float2 &b)
{return make_float2(a.x*b.x, a.y*b.y);}
//Ïðîèçâåäåíèå íà ñêàëÿð:
__device__ float4 operator*(const float4 &a, const float &b)
{return make_float4(a.x*b, a.y*b, a.z*b, a.w*b);}
__device__ float3 operator*(const float3 &a, const float &b)
{return make_float3(a.x*b, a.y*b, a.z*b);}
__device__ float2 operator*(const float2 &a, const float &b)
{return make_float2(a.x*b, a.y*b);}
__device__ float4 operator*(const float &a, const float4 &b)
{return make_float4(b.x*a, b.y*a, b.z*a, b.w*a);}
__device__ float3 operator*(const float &a, const float3 &b)
{return make_float3(b.x*a, b.y*a, b.z*a);}
__device__ float2 operator*(const float &a, const float2 &b)
{return make_float2(b.x*a, b.y*a);}
//Äåëåíèå íà ñêàëÿð:
template <typename T>
__device__ T operator/(const T &a, const float &b)
{return (a*(1.0/b));}
//Íîðìàëèçàöèÿ:
template <typename T>
__device__ T normalize(const T &a)
{return (a/Length(a));}
//typedef float2 Vector2f;
#endif // #ifndef _VECTOR_CUH_
Эта библитека работает в одном треде, но возможно многотредная библиотка окажется выгодней. По затратам на память наверняка, но возможно и по скорости.