2.Metal Shader language (着色语言规范)

2021-02-20  本文已影响0人  Mjs

Metal简述

Metal语言的限制

Metal 基本数据类型

基本数据类型主要有 ：

标量

Metal中的标量类型如下图所示

image 
bool a = true;
char b = 5;
int  d = 15;
//用于表示内存空间
size_t c = 1;
ptrdiff_t f = 2;

向量

向量支持如下类型

//直接赋值初始化
bool2 A= {1,2};
//通过内建函数float4初始化
float4 pos = float4(1.0,2.0,3.0,4.0);

//通过下标从向量中获取某个值
float x = pos[0];
float y = pos[1];

//通过for循环对一个向量进行运算
float4 VB;
for(int i = 0; i < 4 ; i++)
{
    VB[i] = pos[i] * 2.0f;
}

在OpenGL ES的GLSL语言中，例如2.0f，在着色器中书写时，是不能加f，写成2.0，而在Metal中则可以写成2.0f，其中f可以是大写，也可以是小写

向量的访问规则

int4 test = int4(0,1,2,3);
int a = test.x; //获取的向量元素0
int b = test.y; //获取的向量元素1
int c = test.z; //获取的向量元素2
int d = test.w; //获取的向量元素3

int e = test.r; //获取的向量元素0
int f = test.g; //获取的向量元素1
int g = test.b; //获取的向量元素2
int h = test.a; //获取的向量元素3


float4 c;
c.xyzw = float4(1.0f,2.0f,3.0f,4.0f);
c.z = 1.0f;
c.xy = float2(3.0f,4.0f);
c.xyz = float3(3.0f,4.0f,5.0f);


float4 pos = float4(1.0f,2.0f,3.0f,4.0f);
//向量分量逆序访问
float4 swiz = pos.wxyz;  //swiz = (4.0,1.0,2.0,3.0);
//向量分量重复访问
float4 dup = pos.xxyy;  //dup = (1.0f,1.0f,2.0f,2.0f);

//可以仅对 xw / wx 修改
//pos = (5.0f,2.0,3.0,6.0)
pos.xw = float2(5.0f,6.0f);

//pos = (8.0f,2.0f,3.0f,7.0f)
pos.wx = float2(7.0f,8.0f);

//可以仅对 xyz 进行修改
//pos = (3.0f,5.0f,9.0f,7.0f);
pos.xyz = float3(3.0f,5.0f,9.0f);

float2 pos;
pos.x = 1.0f; //合法
pos.z = 1.0f; //非法，pos是二维向量，没有z这个索引

float3 pos2;
pos2.z = 1.0f; //合法
pos2.w = 1.0f; //非法

// 赋值 时 分量不可重复，取值 时 分量可重复
//非法,x出现2次
pos.xx = float2(3.0,4.0f);
pos.xy = swiz.xx;

//向量中xyzw与rgba两组分量不能混合使用
float4 pos4 = float4(1.0f,2.0f,3.0f,4.0f);
pos4.x = 1.0f;
pos4.y = 2.0f;
//非法,.rgba与.xyzw 混合使用
pos4.xg = float2(2.0f,3.0f);
////非法,.rgba与.xyzw 混合使用
float3 coord = pos4.ryz;

GLSL中向量不能乱序访问，只是和Metal中的向量相似，并不是等价。

矩阵

矩阵支持如下类型

float4x4 m;
//将第二行的所有值都设置为2.0
m[1] = float4(2.0f);

//设置第一行/第一列为1.0f
m[0][0] = 1.0f;

//设置第三行第四列的元素为3.0f
m[2][3] = 3.0f;


//float4类型向量的所有可能构造方式
//1个一维向量，表示一行都是x
float4(float x);/
//4个一维向量 --> 4维向量
float4(float x,float y,float z,float w);
//2个二维向量 --> 4维向量
float4(float2 a,float2 b);
//1个二维向量+2个一维向量 --> 4维向量
float4(float2 a,float b,float c);
float4(float a,float2 b,float c);
float4(float a,float b,float2 c);
//1个三维向量+1个一维向量 --> 4维向量
float4(float3 a,float b);
float4(float a,float3 b);
//1个四维向量 --> 4维向量
float4(float4 x);


//float3类型向量的所有可能的构造的方式
//1个一维向量
float3(float x);
//3个一维向量
float3(float x,float y,float z);
//1个一维向量 + 1个二维向量
float3(float a,float2 b);
//1个二维向量 + 1个一维向量
float3(float2 a,float b);
//1个三维向量
float3(float3 x);


//float2类型向量的所有可能的构造方式
//1个一维向量
float2(float x);
//2个一维向量
float2(float x,float y);
//1个二维向量
float2(float2 x);

Metal 其他类型

主要有以下两种

纹理类型

纹理类型是一个句柄，指向一维/二维/三维纹理数据，而纹理数据对应一个纹理的某个level的mipmap的全部或者一部分

纹理的访问权限
在一个函数中描述纹理对象的类型
access枚举值由Metal定义，定义了纹理的访问权利 enum class access {sample, read, write};，有以下3种访问权利，当没写access时，默认的access 就是 sample

定义纹理类型
描述一个纹理对象/类型，有以下三种方式，分别对应一维/二维/三维，

//类型 变量 修饰符
/*
 类型
    - texture2d<float>，读取的数据类型是float，没写access，默认是sample
    - texture2d<float,access::read>，读取的数据类型是float，读取的方式是read
    - texture2d<float,access::write>，读取的数据类型是float，读取的方式是write
 变量名
    - imgA
    - imgB
    - imgC
 修饰符
    - [[texture(0)]] 对应纹理0
    - [[texture(1)]] 对应纹理1
    - [[texture(2)]] 对应纹理2
 */
void foo (texture2d<float> imgA[[texture(0)]],
          texture2d<float,access::read> imgB[[texture(1)]],
          texture2d<float,access::write> imgC[[texture(2)]])
{

    //...
}

采样器类型 Samplers

采样器类型决定了如何对一个纹理进行采样操作，在Metal框架中有一个对应着色器语言的采样器的对象MTLSamplerState，这个对象作为图形渲染着色器函数参数 或是 并行计算函数的参数传递，有以下几种状态：

采样器所有状态如图所示

image

1、openGL ES中纹理坐标对应的是stq，Metal中纹理坐标对应是str
2、在Metal程序中初始化的采样器必须使用constexpr修饰符声明

/*
constexpr：修饰符（必须写）
sampler：类型
s：采样器变量名称
参数
    - coord: 是否需要归一化，不需要归一化，用的是像素pixel
    - address: 地址环绕方式
    - filter: 过滤方式
*/
constexpr sampler s(coord::pixel, address::clamp_to_zero, filter::linear);

constexpr sampler a(coord::normalized);

constexpr sampler b(address::repeat);

函数修饰符

Metal有以下3中函数修饰符，放在函数的最前面，即位于函数返回值的前面

1、使用kernel修饰的函数，其返回值类型必须是void类型
2、一个被函数修饰符修饰的函数不能在调用其他也被函数修饰符修饰的函数，这样会导致编译失败，即Kernel、vertex、fragment修饰的函数不能相互调用，也不能同修饰符函数相互调用。但是可以调用普通函数
3、被函数修饰符修饰过的函数,只允许在客户端对其进行操作. 不允许被普通的函数调用
4、Metal中并不是所有函数都需要上述3个修饰符修饰，是可以在Metal中定义普通函数的，即不带任何修饰符的函数
5、只有图形着色函数才可以被vertex和fragment修饰，对于图形着色函数，通过返回值类型可以辨认出是为顶点计算还是像素计算，其返回值也可以是void，意味着不产生数据输出到绘制管线，是一个无意义的动作

//并行计算函数(kernel)
kernel void CCTestKernelFunctionA(int a,int b)
{ 
    /*
     注意:
     1\. 使用kernel 修饰的函数返回值必须是void 类型
     2\. 一个被函数修饰符修饰过的函数,不允许在调用其他的被函数修饰过的函数. 非法
     3\. 被函数修饰符修饰过的函数,只允许在客户端对其进行操作. 不允许被普通的函数调用.
     */

    //不可以的!
    //一个被函数修饰符修饰过的函数,不允许在调用其他的被函数修饰过的函数. 非法
    CCTestKernelFunctionB(1,2);//非法，错误调用！！！
    CCTestVertexFunctionB(1,2);//非法，错误调用！！！

    //可以! 你可以调用普通函数.而且在Metal 不仅仅只有这3种被修饰过的函数.普通函数也可以存在
    CCTest();

}

//并行计算函数
kernel void CCTestKernelFunctionB(int a,int b)
{
    .....
}

//顶点函数
vertex int CCTestVertexFunctionB(int a,int b)
{
    .....
}

//片元函数
fragment int CCTestVertexFunctionB(int a,int b)
{
    .....
}

//普通函数
void CCTest()
{
    .....
}

变量、参数的地址空间修饰符

Metal着色器语言使用地址空间修饰符来表示一个 函数变量 或者 参数变量 被分配于哪一片内存区域，有以下4中地址空间修饰符

1、所有的着色函数（vertex、fragment、kernel）的参数，如果是指针/引用，都必须带有地址空间修饰符号
2、对于图形着色器函数（即vertex/fragment修饰的函数），其指针/引用类型的参数必须定义为 device、constant地址空间
3、对于并行计算函数（即kernel修饰的函数），其指针/引用类型的参数必须定义为 device、threadgroup、constant
4、并不是所有的变量都需要修饰符，也可以定义普通变量（即无修饰符的变量）

/*
 注意:
 1\. 所有被(kernel,vertex,fragment)所修饰的参数变量,如果其类型是指针/引用 都必须带有地址空间修饰符.
 2\. 被fragment修饰的片元函数, 指针/引用必须被device/constant/threadgroup
 */

//变量/参数地址空间修饰符
void CCTestFouncitionE(device int *g_data,
                       threadgroup int *l_data,
                       constant float *c_data
                       )
{
    //...

}

device：设备地址空间修饰符

除了可以修饰 图形着色器函数 / 并行计算函数参数，还可以修饰指针变量 和 结构体指针变量

// 设备地址空间: device 用来修饰指针.引用
//1.修饰指针变量
device float4 *color;

struct CCStruct{
    float a[3];
    int b[2];
};
//2.修饰结构体类的指针变量
device CCStruct *my_CS;

1、纹理对象总是在设备地址空间分配内存，即纹理对象默认在GPU分配内存
2、device地址空间修饰符不必出现在纹理类型定义中
3、一个纹理对象的内容无法直接访问，Metal提供读写纹理的内建函数，通过内建函数访问纹理对象

threadgroup：线程组地址空间修饰符

/*
 1\. threadgroup 被并行计算计算分配内存变量, 这些变量被一个线程组的所有线程共享. 在线程组分配变量不能被用于图像绘制.
 2\. thread 指向每个线程准备的地址空间. 在其他线程是不可见切不可用的
 */
kernel void CCTestFouncitionF(threadgroup float *a)
{
    //在线程组地址空间分配一个浮点类型变量x
    threadgroup float x;

    //在线程组地址空间分配一个10个浮点类型数的数组y;
    threadgroup float y[10];

}

constant：常量地址空间修饰符

常量地址空间的指针/引用可以作为函数的参数，向声明为常量的变量赋值会产生编译错误
2、声明常量但是没有赋予初值也会产生编译错误

constant float samples[] = { 1.0f, 2.0f, 3.0f, 4.0f };

//对一个常量地址空间的变量进行修改也会失败,因为它只读的
sampler[4] = {3,3,3,3}; //编译失败; 

//定义为常量地址空间声明时不赋初值也会编译失败
constant float a;

thread：线程地址空间修饰符

kernel void CCTestFouncitionG(void)
{
    //在线程空间分配空间给x,p
    float x;
    thread float p = &x;
}

函数参数与变量的传递修饰符，即属性修饰符

图形绘制 或者 并行计算着色器函数的输入输出都是通过参数传递，除了常量地址空间变量和程序域定义的采样器之外, 其他参数修饰的可以是如下之一，有以下5种属性修饰符：

为什么需要属性修饰符？

传递修饰符在代码中的体现
对于每个着色函数来说，一个修饰符是必须指定的，它用来设置一个缓存、纹理、采样器的位置，传递修饰符对应的写法如下：

在代码中的表现如下：

在代码中如何表现:
 1.已知条件:device buffer(设备缓存)/constant buffer(常量缓存)
 代码表现:[[buffer(index)]]
 解读:不变的buffer ,index 可以由开发者来指定.

 2.已知条件:texture Object(纹理对象)
 代码表现: [[texture(index)]]
 解读:不变的texture ,index 可以由开发者来指定.

 3.已知条件:sampler Object(采样器对象)
 代码表示: [[sampler(index)]]
 解读:不变的sampler ,index 可以由开发者来指定.

 4.已知条件:threadgroup Object(线程组对象)
 代码表示: [[threadgroup(index)]]
 解读:不变的threadgroup ,index 可以由开发者来指定.

1、index是一个unsigned interger类型的值，表示了一个缓存、纹理、采样器参数的位置（即在函数参数索引表中的位置，相当于OpenGl ES中的location）
2、从语法上来说，属性修饰符的声明位置应该位于参数变量名之后

//并行计算着色器函数add_vectros ,实现2个设备地址空间中的缓存A与缓存B相加.然后将结果写入到缓存out.
//属性修饰符"(buffer(index))" 为着色函数参数设定了缓存的位置
//thread_position_in_grid：用于表示当前节点在多线程网格中的位置,并不需要开发者传递，是Metal自带的。
/*
 kernel：并行计算函数修饰符
 void：函数返回值类型
 add_vectros：函数名
 const device float4 *inA [[buffer(0)]]：定义了一个float4类型的指针，指向一个4维向量空间，放在设备内存空间（即显存GPU中）
    - const device：只决定放在哪里
    - inA：变量名
    - [[buffer(0)]] 对应 buffer中0这个id
 */
kernel void add_vectros(
                const device float4 *inA [[buffer(0)]],
                const device float4 *inB [[buffer(1)]],
                device float4 *out [[buffer(2)]],
                uint id[[thread_position_in_grid]])
{
    out[id] = inA[id] + inB[id];
}

//着色函数的多个参数使用不同类型的属性修饰符的情况
//纹理读取的方式的sampler，即采样器，[[sampler(0)]]表示采样器的缓存id
kernel void my_kernel(device float4 *p [[buffer(0)]],
                      texture2d<float> img [[texture(0)]],
                      sampler sam [[sampler(0)]])
{
    //.....

}

常见的内建变量修饰符

//定义了片元输入的结构体，
struct MyFragmentOutput {
      // color attachment 0 颜色附着点0
     float4 clr_f [[color(0)]]; 
     // color attachment 1 颜色附着点1
     int4 clr_i [[color(1)]]; 
     // color attachment 2 颜色附着点2
     uint4 clr_ui [[color(2)]]; 
};

fragment MyFragmentOutput my_frag_shader( ... ) 
{
    MyFragmentOutput f;
    ....
    f.clr_f = ...;
    ....
    return f; 
}
上一篇下一篇

猜你喜欢

热点阅读