本文将带你了解IOS开发入门Metal API随着iOS 10与macOS 10.12新引入的Tessellation特性，希望本文对大家学IOS有所帮助。

Tessellation——中文一般译作“细分曲面”，一般用于将由少量顶点构成的面生成细节度更高的面。这其中的原理是将一个三角形或四边形，由GPU根据我们编程的控制点生成规则，自动生成更多的顶点，然后将这些顶点根据一定规则生成更多的三角形。这么一来，我们可以在3D游戏中在远处的敌人使用低模也能做出精细度较高的模型出来了，而且也省顶点数据传输带宽。

在Metal API中，通过tessellation绘制出的图形所走的渲染流水线会与通过传统的顶点着色器所走的渲染流水线会有所不同。Metal API为了简化本身Metal Shading Language的语法，将已有的Compute Shader用来计算控制点生成规则，然后将结果送给Tesselator(曲面细分器)生成具体的控制点，最后交给用于细分曲面后处理的Vertex Shader做顶点数据输出，然后后面的就跟典型的渲染流水线一样处理了。

上图中，蓝色圆角矩形部分是可编程着色器;绿色圆角矩形部分是GPU的固定硬件单元;灰色部分是数据缓存对象。我们可以看到，走Tessellation流水线相对于传统的渲染流水线，其实也就多了计算着色器(用于生成每个patch的细分曲面因子)以及曲面细分器单元。所谓patch其实与一般的三角形(triangle)与四边形(quad)差不多，只是在用于细分曲面的处理过程中，一个patch是在一个平面里的，所以它没有z坐标，我们在细分曲面后处理的顶点着色器中可以给转换后的顶点再做视图模型变换以及投影变换，然后输出。在Metal API中，细分曲面只支持三角形与四边形，不支持线段;在OpenGL中还支持线段。

对于四边形而言，patch的顶点排列与普通的矩形会有些差别。我们通常用Metal API绘制一个矩形往往会用triangle strip的绘制模式(由于Metal API中没有triangle fan的绘制模式)，因此一个矩形的顶点排列可以是：左上，左下，右上，右下这种排列方式。然而绘制patch则不能用这种方式，我们只能对四个顶点依次以顺时针或逆时针的顺序进行排列。四边形patch的控制点位于四边形的边缘以及四边形的内部，因此用于描述四边形patch控制点的结构体定义如下：

typedefstruct{/*NOTE:edgeTessellationFactorandinsideTessellationFactorareinterpretedashalf(16-bitfloats)*/

uint16_tedgeTessellationFactor[4];uint16_tinsideTessellationFactor[2];

}MTLQuadTessellationFactorsHalf; 该结构体中的edgeTessellationFactor成员用于描述四边形四个边缘的控制点生成情况;insideTessellationFactor用于描述内部控制点的生成情况。patch顶点的排列不同，对于edgeTessellationFactor数组对象的各个元素所对应的哪条边也是有所不同的。本文中，我们编排四边形patch的顶点顺序依次为四边形的左下顶点、右下顶点、右上顶点、左上顶点，以逆时针的次序进行编排。这样，edgeTessellationFactor[0]对应的是左边，edgeTessellationFactor[1]对应的是下边，edgeTessellationFactor[2]对应的是右边，edgeTessellationFactor[3]对应的是上边。请见此图：点击看图[objc] view plain copy

typedefstruct{/*NOTE:edgeTessellationFactorandinsideTessellationFactorareinterpretedashalf(16-bitfloats)*/

uint16_tedgeTessellationFactor[3];uint16_tinsideTessellationFactor;

}MTLTriangleTessellationFactorsHalf;

为了统一描述，本文将三角形patch的顶点编排次序排列为：右下、中上、左下。这样使得这里的edgeTessellationFactor[0]对应于三角形的左边，edgeTessellationFactor[1]对应于三角形的下边，edgeTessellationFactor[2]对英语三角形的右边。我们见此图所示：点击看图

三角形边缘上控制点的生成与四边形的类似，而内部控制点生成规则就要复杂不少了。对于insideTessellationFactor值，如果它小于3，那么在三角形内就只有一个控制点，坐落于三角形的重心(所谓三角形重心，即三角形的三个顶点到对边上的中线所交汇的点)处。如果是3，那么在三角形内部则正好有3个控制点构成一个小三角形。如果是4，那么在小三角形中再增加一个控制点，坐落于小三角形的重心处，并且小三角形的每个边缘的中点处增加一个控制点。如果是5，内部小三角形的每条边缘则含有两个控制点，均分边缘;并且在小三角形内部再次构成一个小三角形。如果是6，小三角形的边缘含有3个控制点均分边缘;小三角形内部的小三角形的每个边缘上增加1个控制点，并且在其内部再增加一个控制点，以此类推……具体算法以及相关的图我们可以参考OpenGL官网wiki上关于Tessellation的介绍：https://www.opengl.org/wiki/Tessellation。

三角形patch的每个控制点的坐标以(u, v, w)进行描述，u、v和w的取值均为[0.0, 1.0]区间内。以本文的三角形patch顶点编排顺序为例，三角形左边的所有控制点的u坐标为0;三角形下边上的所有控制点的v坐标为0，三角形右边上所有控制点的w坐标为0。那么在三角形内部的某一个控制点的坐标如何计算得到呢?这里比较晕乎的是在一个二维平面中用三个坐标值来表示一个点，不过这里有意思的一点是：对于三角形内的任一控制点的(u, v, w)坐标，u + v + w = 1.0。所以它们之间是具有关联性的。我们可以从一个简单的例子进行着手分析——对于一个等边三角形，我们假设其中线(对于等边三角形，三线合一，即中线、垂直线、角平分线，另外中心与重心也是同一个)长度为1.0，那么其重心到三条边中点的距离正好都是1/3，那么我们知道，如果将位于重心的控制点的坐标定义为(1/3, 1/3, 1/3)，那么u + v + w的值正好为1.0。所以我们判定三角形patch内部某个控制点的坐标可以这么做：对于判定u值，将三角形左边沿着其中线进行平移，直到与控制点重合，然后看平移后的左边与左边中线的相交点，该点在此中线上的位置即为u值;判断v值也类似，将三角形下边沿着其中线进行平移，然后正好与该控制点重合处，看该下边与其中线相交点在此中线上的位置;w点则是看右边沿着其中线平移的情况。

有了以上对于细分曲面控制点生成的相关概念之后，下面我们就开始描述Metal API中完成细分曲面渲染的整个过程了。

我们首先看如何用计算着色器设置控制点的生成规则。计算着色器的实现如下所示：

constantint4tessOutFactors[[function_constant(0)]];constantint2tessInnerFactors[[function_constant(1)]];

//三角形用于生成控制点的计算内核kernelvoidtriangle_kernel(devicestruct

MTLTriangleTessellationFactorsHalf*factors[[buffer(0)]],uinttid[[thread_position_in_grid]])

{//Simplepassthroughoperation

//Moresophisticatedcomputekernelsmightdeterminethetessellationfactorsbasedonthestateofthescene(e.g.cameradistance)factors[tid].edgeTessellationFactor[0]=tessOutFactors.x;

factors[tid].edgeTessellationFactor[1]=tessOutFactors.y;factors[tid].edgeTessellationFactor[2]=tessOutFactors.z;

factors[tid].insideTessellationFactor=tessInnerFactors.x;}

//四边形用于生成控制点的计算内核

kernelvoidquad_kernel(devicestructMTLQuadTessellationFactorsHalf*factors[[buffer(0)]],

uinttid[[thread_position_in_grid]]){

//Simplepassthroughoperation//Moresophisticatedcomputekernelsmightdeterminethetessellationfactorsbasedonthestateofthescene(e.g.cameradistance)

factors[tid].edgeTessellationFactor[0]=tessOutFactors.x;factors[tid].edgeTessellationFactor[1]=tessOutFactors.y;

factors[tid].edgeTessellationFactor[2]=tessOutFactors.z;factors[tid].edgeTessellationFactor[3]=tessOutFactors.w;

factors[tid].insideTessellationFactor[0]=tessInnerFactors.x;factors[tid].insideTessellationFactor[1]=tessInnerFactors.y;

}

上述Metal Shading Language片段中，最上面的两个常量值将由主机端传入，tessOutFactors是一个四元素向量，每个分量值对应edgeTessellationFactor的相应元素值，这一点我们在两个内核函数中也能看得很清楚。而tessInnerFactors则是一个二元素向量，对于四边形，其第一个分量作为insideTessellationFactor的第0个元素值，第二个分量作为insideTessellationFactor的第1个元素值;对于三角形，tessInnerFactors的第一个分量作为insideTessellationFactor的值。内核函数的参数就是我们上文提到过的指向MTLTriangleTessellationFactorsHalf结构体或MTLQuadTessellationFactorsHalf结构体数组的指针。每一个此结构体对象表征了对某一个patch进行控制点生成的规则，如果我们有多个patch需要做曲面细分，那么这里将会有多个元素，如果就对一个patch进行曲面细分，那么这里其实也就一个元素，我们可以在主机端进行控制给计算内核调度多少个线程进行执行。

以下代码片段是主机端对计算着色器的相关设置：

//0号对应外部边缘的tess。对于四边形，顺序为左、下、右、上;对于三角形，顺序为：左、下、右[constantValuessetConstantValue:(constint[]){5,4,3,2}type:MTLDataTypeInt4atIndex:0];

//1号对应内部的tess。对于四边形，分别为水平方向tess、垂直方向tess;三角形只能用一个值

[constantValuessetConstantValue:(constint[]){4,3}type:MTLDataTypeInt2atIndex:1];

computeProgram=[mLibrarynewFunctionWithName:tessControlKernelNameconstantValues:constantValueserror:NULL];if(computeProgram==nil)

{NSLog(@"计算内核获取失败");

break;}

[constantValuesrelease];

mComputePipelineState=[devicenewComputePipelineStateWithFunction:computeProgramerror:NULL];

if(mComputePipelineState==nil){

NSLog(@"计算流水线创建失败");break;

}

//由于细分曲面因子缓存最终由生成控制点的内核程序生成，因此不会被CPU读写，而是直接在流水线内部使用，因此这里使用GPU私有存储模式mTessFactorsBuffer=[devicenewBufferWithLength:256options:MTLResourceStorageModePrivate];

上述代码片段设置了上面所提到的Metal Shading Language中的常量向量对象，另外创建了计算着色器程序，并建立了计算流水线。mTessFactorsBuffer这个缓存对象中就将存放着色器中的MTLTriangleTessellationFactorsHalf结构体或MTLQuadTessellationFactorsHalf结构体的数据。由于我们这里仅仅对一个patch进行绘制，所以就假定其长度为256个字节，这对于存放这俩结构体的内容是绰绰有余了。

下面我们看看如何调度执行计算着色器。

idcommandBuffer=[mCommandQueuecommandBuffer];

//设置命令缓存执行完成后的处理[commandBufferaddCompletedHandler:^void(idcmdBuf){

//命令全都执行完之后，将mCurrentDrawable置空，表示可以绘制下面一帧mCurrentDrawable=nil;

}];

//从命令缓存对象获取计算命令编码器idcomputeCommandEncoder=[commandBuffercomputeCommandEncoder];

//设置计算流水线

[computeCommandEncodersetComputePipelineState:mComputePipelineState];

//设置内核程序的缓存参数[computeCommandEncodersetBuffer:mTessFactorsBufferoffset:0atIndex:0];

//分派线程组

//由于我们这里仅对一个patch做控制点生成规则，因此只需要一个线程进行处理即可[computeCommandEncoderdispatchThreadgroups:MTLSizeMake(1,1,1)threadsPerThreadgroup:MTLSizeMake(1,1,1)];

[computeCommandEncoderendEncoding];

上述代码片段首先获得了命令缓存对象，然后通过该命令缓存对象获取计算命令编码器，对它设置计算流水线以及相关缓存对象。然后我们用计算命令编码器对计算内核进行分派，这样就激活了计算着色器的计算执行，然后结束命令编码器。在Metal中，一个命令缓存对象可以有多个命令编码器对其进行操作，一般情况下，同一时段只能由一个命令编码器处于活动状态，然后一直到它结束编码为止，另一个命令编码器才能操作。等到所有命令编码器完成操作之后，命令缓存可以做提交操作，让GPU做真正的命令执行。

这里对于计算着色器而言，MTLTriangleTessellationFactorsHalf结构体或MTLQuadTessellationFactorsHalf结构体数组其实是作为输出，并且该结构体类型是固定的，我们不要用其他类型来代替它，不过我们可以再添加其他缓存对象来输入一些必要的参数。

控制点生成规则数据创建完之后，它是怎么被传到细分曲面器tessellator中的呢?首先，计算着色器计算好的MTLTriangleTessellationFactorsHalf结构体或MTLQuadTessellationFactorsHalf结构体的数据都是直接存入主机端所创建的mTessFactorsBuffer缓存对象中的。然后，我们将mTessFactorsBuffer缓存对象传递给渲染命令编码器，最后通过渲染编码器调用drawPatches接口来激活tessellator，并执行后续的渲染流水线。

我们在看细分曲面后处理顶点着色器之前，先看一下主机端的一些设置。首先，我们看一下我们所指定的四边形与三角形patch的顶点排列：

//以下以逆时针方向构成一个四边形，使得四边形的边顺序依次为左、下、右、上。staticconstfloatsRectangleVertices[]={

//左下顶点-0.9f,-0.9f,0.9f,0.1f,0.1f,1.0f,

//右下顶点

0.9f,-0.9f,0.1f,0.9f,0.1f,1.0f,

//右上顶点0.9f,0.9f,0.1f,0.1f,0.9f,1.0f,

//左上顶点

-0.9f,0.9f,0.9f,0.9f,0.1f,1.0f,};

//以下以逆时针方向构成一个三角形，使得三角形的边顺序依次为左、下、右。

staticconstfloatsTriangleVertices[]={

//右下顶点0.9f,-0.9f,0.1f,0.1f,0.9f,1.0f,

//中上顶点

0.0f,0.9f,0.9f,0.1f,0.1f,1.0f,

//左下顶点-0.9f,-0.9f,0.1f,0.9f,0.1f,1.0f,

};

这两个patch的顶点分布与我们一开始所提到的顶点分布顺序一致。各位如果看Apple官方的《The Metal Programming Guide》中的例子的话，其中那张四边形patch的图是以：左上、右上、右下、左下的顺时针次序排列而得到的。而我们这里都采用逆时针排列方式。下面是对渲染流水线相关的设置：

vertexProgram=[mLibrarynewFunctionWithName:tessEvalVertexName];if(vertexProgram==nil)

{NSLog(@"顶点着色器获取失败");

break;}

fragmentProgram=[mLibrarynewFunctionWithName:@"square_fragment"];

if(fragmentProgram==nil){

NSLog(@"片段着色器获取失败");break;

}

MTLVertexDescriptor*vertexDescriptor=[MTLVertexDescriptornew];vertexDescriptor.attributes[0].format=MTLVertexFormatFloat2;

vertexDescriptor.attributes[0].bufferIndex=0;//顶点结构体的此属性对应的buffer索引为0vertexDescriptor.attributes[0].offset=0;

vertexDescriptor.attributes[1].format=MTLVertexFormatFloat4;vertexDescriptor.attributes[1].bufferIndex=0;//顶点结构体的此属性对应的buffer索引为0

vertexDescriptor.attributes[1].offset=8;//第二个属性color偏移是从第8字节开始

vertexDescriptor.layouts[0].stride=6*sizeof(float);vertexDescriptor.layouts[0].stepFunction=MTLVertexStepFunctionPerPatchControlPoint;

vertexDescriptor.layouts[0].stepRate=1;

MTLRenderPipelineDescriptor*descriptor=[MTLRenderPipelineDescriptornew];

//Configurecommonrenderpropertiesdescriptor.sampleCount=4;//我们将使用多重采样抗锯齿(MSAA)，每个像素由4个样本构成

descriptor.vertexFunction=vertexProgram;descriptor.fragmentFunction=fragmentProgram;

descriptor.depthAttachmentPixelFormat=MTLPixelFormatInvalid;//不启用深度测试descriptor.stencilAttachmentPixelFormat=MTLPixelFormatInvalid;//不启用模版测试

descriptor.vertexDescriptor=vertexDescriptor;[vertexDescriptorrelease];

descriptor.colorAttachments[0].pixelFormat=self.pixelFormat;

descriptor.colorAttachments[0].blendingEnabled=YES;//将渲染流水线设置为允许颜色混合descriptor.colorAttachments[0].rgbBlendOperation=MTLBlendOperationAdd;

descriptor.colorAttachments[0].alphaBlendOperation=MTLBlendOperationAdd;descriptor.colorAttachments[0].sourceRGBBlendFactor=MTLBlendFactorSourceAlpha;

descriptor.colorAttachments[0].sourceAlphaBlendFactor=MTLBlendFactorSourceAlpha;descriptor.colorAttachments[0].destinationRGBBlendFactor=MTLBlendFactorOneMinusSourceAlpha;

descriptor.colorAttachments[0].destinationAlphaBlendFactor=MTLBlendFactorOneMinusSourceAlpha;

//配置细分曲面常用属性descriptor.tessellationFactorScaleEnabled=NO;

descriptor.tessellationFactorFormat=MTLTessellationFactorFormatHalf;descriptor.tessellationControlPointIndexType=MTLTessellationControlPointIndexTypeNone;

descriptor.tessellationFactorStepFunction=MTLTessellationFactorStepFunctionConstant;//由于顶点构成是以逆时针作为顺序的，因此这里使用逆时针方向

descriptor.tessellationOutputWindingOrder=MTLWindingCounterClockwise;descriptor.tessellationPartitionMode=MTLTessellationPartitionModeInteger;

//在macOS中，最大细分曲面因子为64，默认为16

descriptor.maxTessellationFactor=64;

mRenderPipelineState=[devicenewRenderPipelineStateWithDescriptor:descriptorerror:NULL];[descriptorrelease];

if(mRenderPipelineState==nil){

NSLog(@"渲染流水线创建失败");break;

}

if(mIsForRectangle)mVertexBuffer=[devicenewBufferWithBytes:sRectangleVerticeslength:sizeof(sRectangleVertices)options:MTLResourceCPUCacheModeWriteCombined];

elsemVertexBuffer=[devicenewBufferWithBytes:sTriangleVerticeslength:sizeof(sTriangleVertices)options:MTLResourceCPUCacheModeWriteCombined];

各位这里要注意的是，对于细分曲面后处理的顶点着色器，原patch的顶点必须以stage-in的方式传入，而不能是一个普通的数据缓存对象。此外，在Metal Shading Language中，我们必须用patch_control_point结构体模板对输入的patch结构体进行封装。细分曲面后处理顶点着色器函数逐个对patch顶点数据进行读入，而我们这里必须使用patch_control_point结构体对象作为输入参数。此外，我们看到上述顶点描述符中，其layouts[0]的stepFunction成员必须用MTLVertexStepFunctionPerPatchControlPoint，表示该顶点将作为patch控制点来使用。关于tessellationPartitionMode模式的详细介绍，各位可以参考https://www.cnblogs.com/zenny-chen/p/4280100.html此博文中“指定细分曲面坐标的空间”部分，这里使用MTLTessellationPartitionModeInteger表示均分。创建patch顶点缓存的方式与创建普通图形顶点缓存的方式一样。

下面我们来看一下如何渲染patch。

[objc] view plain copy

//从命令缓存对象获取渲染命令编码器idrenderCommandEncoder=[commandBufferrenderCommandEncoderWithDescriptor:mRenderPassDescriptor];

[renderCommandEncodersetRenderPipelineState:mRenderPipelineState];

[renderCommandEncodersetCullMode:MTLCullModeBack];[renderCommandEncodersetFrontFacingWinding:MTLWindingCounterClockwise];

[renderCommandEncodersetTriangleFillMode:mIsWireFrameMode?MTLTriangleFillModeLines:MTLTriangleFillModeFill];

[renderCommandEncodersetVertexBuffer:mVertexBufferoffset:0atIndex:0];[renderCommandEncodersetTessellationFactorBuffer:mTessFactorsBufferoffset:0instanceStride:0];

constNSUIntegerpatchControlPoints=mIsForRectangle?4:3;

[renderCommandEncoderdrawPatches:patchControlPointspatchStart:0patchCount:1patchIndexBuffer:NULLpatchIndexBufferOffset:0instanceCount:1baseInstance:0];

[renderCommandEncoderendEncoding];

[commandBufferpresentDrawable:mCurrentDrawable];

[commandBuffercommit]; 这里我们调用渲染命令编码器的setTessellationFactorBuffer接口，将我们在计算着色器中输出的存放细分曲面因子数据的mTessFactorBuffer缓存对象传递进去。这个对象不需要在Metal Shader中作为参数传递进去，因为它直接会交给tessellator固定功能单元，而在细分曲面后处理顶点着色器中我们不需要关心这些数据，我们得到的是将是tessellator帮我们生成好的具体控制点信息。这里大家需要注意的是，我们应该先让计算命令编码器活动工作，结束后再使用渲染命令编码器，最后提交命令缓存。

下面给出细分曲面后处理顶点着色器：

view plaincopy

//控制点结构体structControlPoint{

float2position[[attribute(0)]];float4color[[attribute(1)]];

};

//Patch结构体structPatchIn{

patch_control_pointcontrol_points;};

//顶点着色器输出到片段着色器的结构体

structFunctionOutIn{float4position[[position]];

half4color[[flat]];};

//三角形细分曲面后处理顶点着色器

[[patch(triangle,3)]]vertexstructFunctionOutIntriangle_vertex(structPatchInpatchIn[[stage_in]],

float3patch_coord[[position_in_patch]]){

floatu=patch_coord.x;floatv=patch_coord.y;

floatw=patch_coord.z;

//将当前控制点坐标(u,v,w)通过线性插值转换为笛卡尔坐标(x,y)floatx=u*patchIn.control_points[0].position.x+v*patchIn.control_points[1].position.x+w*patchIn.control_points[2].position.x;

floaty=u*patchIn.control_points[0].position.y+v*patchIn.control_points[1].position.y+w*patchIn.control_points[2].position.y;

//顶点输出structFunctionOutInvertexOut;

vertexOut.position=float4(x,y,0.0,1.0);vertexOut.color=half4(u,v,w,1.0);

returnvertexOut;}

//四边形细分曲面后处理顶点着色器

[[patch(quad,4)]]vertexstructFunctionOutInquad_vertex(structPatchInpatchIn[[stage_in]],

float2patch_coord[[position_in_patch]]){

//从tessellator处理之后所获得的规格化之后的控制点坐标——//uv坐标的原点(即(0,0)的位置)是在原patch的左下顶点

floatu=patch_coord.x;floatv=patch_coord.y;

//以下通过线性插值的算法将规格化后的控制点坐标再转换为相对于输入顶点的坐标

float2lower_middle=mix(patchIn.control_points[0].position.xy,patchIn.control_points[1].position.xy,u);float2upper_middle=mix(patchIn.control_points[2].position.xy,patchIn.control_points[3].position.xy,1-u);

//顶点输出

structFunctionOutInvertexOut;vertexOut.position=float4(mix(lower_middle,upper_middle,v),0.0f,1.0f);

vertexOut.color=half4(u,v,1.0f-v,1.0h);

//靠左下的所有顶点使用原patch左下顶点的颜色if(u<0.5f&&v<0.5f)

vertexOut.color=half4(patchIn.control_points[0].color);//靠右下的所有顶点使用原patch右下顶点的颜色

elseif(u>0.5f&&v<0.5f)vertexOut.color=half4(patchIn.control_points[1].color);

//靠右上的所有顶点使用原patch右上顶点的颜色elseif(u>0.5f&&v>0.5f)

vertexOut.color=half4(patchIn.control_points[2].color);//靠左上的所有顶点使用原patch左上顶点的颜色

elseif(u<0.5f&&v>0.5f)vertexOut.color=half4(patchIn.control_points[3].color);

returnvertexOut;

} 上述代码片段中，我们看到，细分曲面后处理顶点着色器通过[[ patch() ]]属性进行修饰来描述，在macOS中，我们必须指定当前patch顶点个数，而在iOS中则可缺省。在细分曲面后处理顶点着色器中，patch_coord参数存放的是来自tessellator的当前控制点的相对坐标(四边形为(u, v)，三角形为(u, v, w))。我们在此着色器中就需要将这些控制点坐标转换为我们标准笛卡尔坐标系下的坐标值，也就是与原patch顶点坐标相一致的坐标系的值。在四边形的后处理着色器中，我们还通过判定控制点所在的区域，分别给它们设置我们在顶点数据中所包含的颜色数据。最后，vertexOut的输出就最终给光栅化器，然后交给片段着色器。

最后，如果此博文有些疏漏或表达不准确的地方也请其他大牛不吝指出，我也会做相应修改，谢谢。

本文由职坐标整理并发布，希望对同学们有所帮助。了解更多详情请关注职坐标移动开发之IOS频道！