Metal shading language pixel format

Question

fireman_play OP

Created Sep ’16

Replies 7

Boosts 0

Views 1.4k

Participants 2

Hi everyone,

I am new to metal and GPU programming and I hope you can help me with the following questions.

I would like to write a GPU kernel functions to process textures. Something like this:

let encoder = commandBuffer.makeComputeCommandEncoder()

encoder.setComputePipelineState(my_kernel_pipeline)

encoder.setTexture(input_texture.texture, at: 0)

encoder.setTexture(output_texture.texture, at: 1)

let threadsPerGroups = MTLSizeMake(8, 8, 1)

...

The pixelFormat of the "input_texture" is rgba8norm (8 bits RGBA color) and the pixelFormat of the "output_texture" is rgba16float (16 bits float). My kernel shaders code is:

kernel void my_kernel(texture2d<***, access::read> inTexture [[texture(0)]],

texture2d<YYY, access::write> outTexture [[texture(1)]],

uint2 gid [[thread_position_in_grid]]) {

....

}

Should *** and YYY be different? What exactly should *** and YYY be?

Thank you very much.

George

Boost

Answer 1

wcm OP

Sep ’16

Reading or sampling from a texture will produce a 4-component vector of whatever type is specified in the template parameter. For example, if

***

were

float

, sampling from

inTexture

would produce a

float4

, where the components have been scaled and biased to the [0,1] range. The same rule applies in reverse when writing.

0

Answer 2

fireman_play OP

Sep ’16

@wcm. thank you very much. So I probably should set *** to "unsigned char" to make it more efficient?

0

Answer 3

fireman_play OP

Sep ’16

@wcm, What I want to do with this kernel funciton is to subtract a constant RGB values from the original color and then convert it to float. For example, if the input color is (200, 100, 100), and the constant is (128, 128, 128), the output should be (72.0, -28.0, -28.0). Should my kernel funciton be this:

float4 inColor = inTexture.read(gid);

float4 outColor = float4(inColor.x*255.0 - 128.0, inColor.y*255.0 - 128.0, inColor.z*255.0 - 128.0, 0.0);

outTexture.write(outColor, gid);

or

float4 inColor = inTexture.read(gid);

float4 outColor = float4(inColor.x - 0.5, inColor.y - 0.5, inColor.z - 0.5, 0.0);

outTexture.write(outColor, gid);

Thank you very much.

0

Answer 4

wcm OP

Sep ’16

The latter. You could also use

half

instead of

float

throughout for slightly more efficient arithmetic.

0

Answer 5

fireman_play OP

Sep ’16

@wcm Thanks, just want to confirm the latter is right even my output texture is in float format?

0

Answer 6

wcm OP

Sep ’16

Yes, whatever format you write to the texture should match the data type as specified in the texture parameter. This is enforced by the compiler.

0

Answer 7

fireman_play OP

Sep ’16

Thanks. But what if inside the kernel function the data becemes larger than 1? for example:

float4 inColor = inTexture.read(gid);

float4 outColor = float4(inColor.x * 2, inColor.y * 2, inColor.z * 2, 0.0);

outTexture.write(outColor, gid);

If the input is (0.8, 0.8, 0.8, 0), the output will be (1.6, 1,6, 1,6, 0) or (1.0, 1.0, 1.0, 0)?

0