Merge branch 'gl_canvas' of ssh://git.forkingpaths.dev/martinfouilleul/milepost into gl_canvas
This commit is contained in:
commit
c982b524c0
|
|
@ -6,6 +6,6 @@ SRCDIR=../../src
|
|||
|
||||
INCLUDES="-I$SRCDIR -I$SRCDIR/util -I$SRCDIR/platform -I$SRCDIR/app"
|
||||
LIBS="-L$BINDIR -lmilepost -framework Carbon -framework Cocoa -framework Metal -framework QuartzCore"
|
||||
FLAGS="-mmacos-version-min=10.15.4 -DDEBUG -DLOG_COMPILE_DEBUG -Wl,-dead_strip"
|
||||
FLAGS="-mmacos-version-min=10.15.4 -DDEBUG -DLOG_COMPILE_DEBUG -Wl,-dead_strip -DMG_IMPLEMENTS_BACKEND_METAL"
|
||||
|
||||
clang -g $FLAGS $LIBS $INCLUDES -o $BINDIR/example_canvas main.c
|
||||
|
|
|
|||
|
|
@ -16,11 +16,44 @@
|
|||
|
||||
#define LOG_SUBSYSTEM "Main"
|
||||
|
||||
mg_font create_font()
|
||||
{
|
||||
//NOTE(martin): create font
|
||||
str8 fontPath = mp_app_get_resource_path(mem_scratch(), "../resources/OpenSansLatinSubset.ttf");
|
||||
char* fontPathCString = str8_to_cstring(mem_scratch(), fontPath);
|
||||
|
||||
FILE* fontFile = fopen(fontPathCString, "r");
|
||||
if(!fontFile)
|
||||
{
|
||||
LOG_ERROR("Could not load font file '%s'\n", fontPathCString);
|
||||
return(mg_font_nil());
|
||||
}
|
||||
unsigned char* fontData = 0;
|
||||
fseek(fontFile, 0, SEEK_END);
|
||||
u32 fontDataSize = ftell(fontFile);
|
||||
rewind(fontFile);
|
||||
fontData = (unsigned char*)malloc(fontDataSize);
|
||||
fread(fontData, 1, fontDataSize, fontFile);
|
||||
fclose(fontFile);
|
||||
|
||||
unicode_range ranges[5] = {UNICODE_RANGE_BASIC_LATIN,
|
||||
UNICODE_RANGE_C1_CONTROLS_AND_LATIN_1_SUPPLEMENT,
|
||||
UNICODE_RANGE_LATIN_EXTENDED_A,
|
||||
UNICODE_RANGE_LATIN_EXTENDED_B,
|
||||
UNICODE_RANGE_SPECIALS};
|
||||
|
||||
mg_font font = mg_font_create_from_memory(fontDataSize, fontData, 5, ranges);
|
||||
free(fontData);
|
||||
|
||||
return(font);
|
||||
}
|
||||
|
||||
int main()
|
||||
{
|
||||
LogLevel(LOG_LEVEL_DEBUG);
|
||||
|
||||
mp_init();
|
||||
mp_clock_init();
|
||||
|
||||
mp_rect rect = {.x = 100, .y = 100, .w = 800, .h = 600};
|
||||
mp_window window = mp_window_create(rect, "test", 0);
|
||||
|
|
@ -29,20 +62,27 @@ int main()
|
|||
#if defined(OS_MACOS)
|
||||
mg_surface surface = mg_metal_surface_create_for_window(window);
|
||||
#elif defined(OS_WIN64)
|
||||
mg_surface surface = mg_gles_surface_create_for_window(window);
|
||||
mg_surface surface = mg_gl_surface_create_for_window(window);
|
||||
#else
|
||||
#error "unsupported OS"
|
||||
#endif
|
||||
|
||||
//TODO: create canvas
|
||||
mg_canvas canvas = mg_canvas_create(surface);
|
||||
mg_font font = create_font();
|
||||
|
||||
// start app
|
||||
mp_window_bring_to_front(window);
|
||||
mp_window_focus(window);
|
||||
|
||||
f32 x = 400, y = 300;
|
||||
f32 dx = 5, dy = 5;
|
||||
f64 frameTime = 0;
|
||||
|
||||
while(!mp_should_quit())
|
||||
{
|
||||
f64 startTime = mp_get_time(MP_CLOCK_MONOTONIC);
|
||||
|
||||
mp_pump_events(0);
|
||||
mp_event event = {0};
|
||||
while(mp_next_event(&event))
|
||||
|
|
@ -63,58 +103,29 @@ int main()
|
|||
event.frame.rect.h);
|
||||
} break;
|
||||
|
||||
case MP_EVENT_WINDOW_MOVE:
|
||||
{
|
||||
printf("moved, rect = {%f, %f, %f, %f}\n",
|
||||
event.frame.rect.x,
|
||||
event.frame.rect.y,
|
||||
event.frame.rect.w,
|
||||
event.frame.rect.h);
|
||||
} break;
|
||||
|
||||
case MP_EVENT_MOUSE_MOVE:
|
||||
{
|
||||
printf("mouse moved, pos = {%f, %f}, delta = {%f, %f}\n",
|
||||
event.move.x,
|
||||
event.move.y,
|
||||
event.move.deltaX,
|
||||
event.move.deltaY);
|
||||
} break;
|
||||
|
||||
case MP_EVENT_MOUSE_WHEEL:
|
||||
{
|
||||
printf("mouse wheel, delta = {%f, %f}\n",
|
||||
event.move.deltaX,
|
||||
event.move.deltaY);
|
||||
} break;
|
||||
|
||||
case MP_EVENT_MOUSE_ENTER:
|
||||
{
|
||||
printf("mouse enter\n");
|
||||
} break;
|
||||
|
||||
case MP_EVENT_MOUSE_LEAVE:
|
||||
{
|
||||
printf("mouse leave\n");
|
||||
} break;
|
||||
|
||||
case MP_EVENT_MOUSE_BUTTON:
|
||||
{
|
||||
printf("mouse button %i: %i\n",
|
||||
event.key.code,
|
||||
event.key.action == MP_KEY_PRESS ? 1 : 0);
|
||||
} break;
|
||||
|
||||
case MP_EVENT_KEYBOARD_KEY:
|
||||
{
|
||||
printf("key %i: %s\n",
|
||||
event.key.code,
|
||||
event.key.action == MP_KEY_PRESS ? "press" : (event.key.action == MP_KEY_RELEASE ? "release" : "repeat"));
|
||||
} break;
|
||||
|
||||
case MP_EVENT_KEYBOARD_CHAR:
|
||||
if(event.key.action == MP_KEY_PRESS || event.key.action == MP_KEY_REPEAT)
|
||||
{
|
||||
printf("entered char %s\n", event.character.sequence);
|
||||
/*
|
||||
if(event.key.code == MP_KEY_LEFT)
|
||||
{
|
||||
dx-=5.1;
|
||||
}
|
||||
else if(event.key.code == MP_KEY_RIGHT)
|
||||
{
|
||||
dx+=5.1;
|
||||
}
|
||||
else if(event.key.code == MP_KEY_UP)
|
||||
{
|
||||
dy+=5.1;
|
||||
}
|
||||
else if(event.key.code == MP_KEY_DOWN)
|
||||
{
|
||||
dy-=5.1;
|
||||
}
|
||||
*/
|
||||
}
|
||||
} break;
|
||||
|
||||
default:
|
||||
|
|
@ -122,29 +133,72 @@ int main()
|
|||
}
|
||||
}
|
||||
|
||||
if(x-200 < 0)
|
||||
{
|
||||
dx = 5;
|
||||
}
|
||||
if(x+200 > 800)
|
||||
{
|
||||
dx = -5;
|
||||
}
|
||||
if(y-200 < 0)
|
||||
{
|
||||
dy = 5;
|
||||
}
|
||||
if(y+200 > 550)
|
||||
{
|
||||
dy = -5;
|
||||
}
|
||||
x += dx;
|
||||
y += dy;
|
||||
|
||||
mg_surface_prepare(surface);
|
||||
|
||||
// background
|
||||
mg_set_color_rgba(1, 0, 1, 1);
|
||||
mg_set_color_rgba(0, 1, 1, 1);
|
||||
mg_clear();
|
||||
|
||||
// head
|
||||
mg_set_color_rgba(1, 1, 0, 1);
|
||||
mg_circle_fill(400, 300, 200);
|
||||
mg_circle_fill(x, y, 200);
|
||||
|
||||
// smile
|
||||
mg_set_color_rgba(1, 0, 0, 1);
|
||||
f32 frown = frameTime > 0.033 ? 100 : 0;
|
||||
|
||||
mg_set_color_rgba(0, 0, 0, 1);
|
||||
|
||||
mg_set_width(20);
|
||||
mg_move_to(300, 200);
|
||||
mg_cubic_to(350, 150, 450, 150, 500, 200);
|
||||
mg_move_to(x-100, y-100);
|
||||
mg_cubic_to(x-50, y-150+frown, x+50, y-150+frown, x+100, y-100);
|
||||
mg_stroke();
|
||||
|
||||
// eyes
|
||||
mg_ellipse_fill(330, 350, 30, 50);
|
||||
mg_ellipse_fill(470, 350, 30, 50);
|
||||
mg_ellipse_fill(x-70, y+50, 30, 50);
|
||||
mg_ellipse_fill(x+70, y+50, 30, 50);
|
||||
|
||||
// text
|
||||
mg_set_color_rgba(0, 0, 1, 1);
|
||||
mg_set_font(font);
|
||||
mg_set_font_size(12);
|
||||
mg_move_to(50, 50);
|
||||
|
||||
str8 text = str8_pushf(mem_scratch(),
|
||||
"Milepost vector graphics test program (frame time = %fs, fps = %f)...",
|
||||
frameTime,
|
||||
1./frameTime);
|
||||
mg_text_outlines(text);
|
||||
mg_fill();
|
||||
|
||||
printf("Milepost vector graphics test program (frame time = %fs, fps = %f)...\n",
|
||||
frameTime,
|
||||
1./frameTime);
|
||||
|
||||
mg_flush();
|
||||
mg_surface_present(surface);
|
||||
|
||||
mem_arena_clear(mem_scratch());
|
||||
frameTime = mp_get_time(MP_CLOCK_MONOTONIC) - startTime;
|
||||
|
||||
}
|
||||
|
||||
mp_terminate();
|
||||
|
|
|
|||
|
|
@ -80,7 +80,6 @@ int main()
|
|||
|
||||
f32 x = 400, y = 300;
|
||||
f32 dx = 5, dy = 5;
|
||||
|
||||
f64 frameTime = 0;
|
||||
|
||||
while(!mp_should_quit())
|
||||
|
|
|
|||
|
|
@ -2605,15 +2605,19 @@ mg_canvas mg_canvas_create(mg_surface surface)
|
|||
break;
|
||||
#endif
|
||||
|
||||
/*
|
||||
#ifdef MG_IMPLEMENTS_BACKEND_GLES
|
||||
case MG_BACKEND_GLES:
|
||||
backend = mg_gles_canvas_create(surface);
|
||||
break;
|
||||
#endif
|
||||
|
||||
*/
|
||||
#ifdef MG_IMPLEMENTS_BACKEND_GL
|
||||
case MG_BACKEND_GL:
|
||||
backend = mg_gl_canvas_create(surface);
|
||||
break;
|
||||
#endif
|
||||
|
||||
default:
|
||||
break;
|
||||
}
|
||||
|
|
|
|||
|
|
@ -25,6 +25,8 @@ typedef struct mg_metal_canvas_backend
|
|||
mg_canvas_backend interface;
|
||||
mg_surface surface;
|
||||
|
||||
mg_color clearColor;
|
||||
|
||||
// permanent metal resources
|
||||
id<MTLComputePipelineState> tilingPipeline;
|
||||
id<MTLComputePipelineState> sortingPipeline;
|
||||
|
|
@ -37,6 +39,7 @@ typedef struct mg_metal_canvas_backend
|
|||
// textures and buffers
|
||||
id<MTLTexture> outTexture;
|
||||
id<MTLTexture> atlasTexture;
|
||||
id<MTLBuffer> shapeBuffer;
|
||||
id<MTLBuffer> vertexBuffer;
|
||||
id<MTLBuffer> indexBuffer;
|
||||
id<MTLBuffer> tileCounters;
|
||||
|
|
@ -57,7 +60,20 @@ mg_metal_surface* mg_metal_canvas_get_surface(mg_metal_canvas_backend* canvas)
|
|||
return(res);
|
||||
}
|
||||
|
||||
void mg_metal_canvas_draw_buffers(mg_canvas_backend* interface, u32 vertexCount, u32 indexCount, mg_color clearColor)
|
||||
void mg_metal_canvas_begin(mg_canvas_backend* interface)
|
||||
{}
|
||||
|
||||
void mg_metal_canvas_end(mg_canvas_backend* interface)
|
||||
{}
|
||||
|
||||
void mg_metal_canvas_clear(mg_canvas_backend* interface, mg_color clearColor)
|
||||
{
|
||||
//TODO
|
||||
mg_metal_canvas_backend* backend = (mg_metal_canvas_backend*)interface;
|
||||
backend->clearColor = clearColor;
|
||||
}
|
||||
|
||||
void mg_metal_canvas_draw_batch(mg_canvas_backend* interface, u32 vertexCount, u32 indexCount, u32 shapeCount)
|
||||
{
|
||||
mg_metal_canvas_backend* backend = (mg_metal_canvas_backend*)interface;
|
||||
mg_metal_surface* surface = mg_metal_canvas_get_surface(backend);
|
||||
|
|
@ -92,9 +108,10 @@ void mg_metal_canvas_draw_buffers(mg_canvas_backend* interface, u32 vertexCount,
|
|||
[boxEncoder setComputePipelineState: backend->boxingPipeline];
|
||||
[boxEncoder setBuffer: backend->vertexBuffer offset:0 atIndex: 0];
|
||||
[boxEncoder setBuffer: backend->indexBuffer offset:0 atIndex: 1];
|
||||
[boxEncoder setBuffer: backend->triangleArray offset:0 atIndex: 2];
|
||||
[boxEncoder setBuffer: backend->boxArray offset:0 atIndex: 3];
|
||||
[boxEncoder setBytes: &scale length: sizeof(float) atIndex: 4];
|
||||
[boxEncoder setBuffer: backend->shapeBuffer offset:0 atIndex: 2];
|
||||
[boxEncoder setBuffer: backend->triangleArray offset:0 atIndex: 3];
|
||||
[boxEncoder setBuffer: backend->boxArray offset:0 atIndex: 4];
|
||||
[boxEncoder setBytes: &scale length: sizeof(float) atIndex: 5];
|
||||
|
||||
MTLSize boxGroupSize = MTLSizeMake(backend->boxingPipeline.maxTotalThreadsPerThreadgroup, 1, 1);
|
||||
MTLSize boxGridSize = MTLSizeMake(indexCount/3, 1, 1);
|
||||
|
|
@ -137,20 +154,21 @@ void mg_metal_canvas_draw_buffers(mg_canvas_backend* interface, u32 vertexCount,
|
|||
[sortEncoder endEncoding];
|
||||
|
||||
//-----------------------------------------------------------
|
||||
//NOTE(martin): create compute encoder and encode commands
|
||||
//NOTE(martin): encode drawing pass
|
||||
//-----------------------------------------------------------
|
||||
vector_float4 clearColorVec4 = {clearColor.r, clearColor.g, clearColor.b, clearColor.a};
|
||||
vector_float4 clearColorVec4 = {backend->clearColor.r, backend->clearColor.g, backend->clearColor.b, backend->clearColor.a};
|
||||
|
||||
id<MTLComputeCommandEncoder> encoder = [surface->commandBuffer computeCommandEncoder];
|
||||
[encoder setComputePipelineState:backend->computePipeline];
|
||||
[encoder setTexture: backend->outTexture atIndex: 0];
|
||||
[encoder setTexture: backend->atlasTexture atIndex: 1];
|
||||
[encoder setBuffer: backend->vertexBuffer offset:0 atIndex: 0];
|
||||
[encoder setBuffer: backend->tileCounters offset:0 atIndex: 1];
|
||||
[encoder setBuffer: backend->tilesArray offset:0 atIndex: 2];
|
||||
[encoder setBuffer: backend->triangleArray offset:0 atIndex: 3];
|
||||
[encoder setBuffer: backend->boxArray offset:0 atIndex: 4];
|
||||
[encoder setBytes: &clearColorVec4 length: sizeof(vector_float4) atIndex: 5];
|
||||
[encoder setBuffer: backend->shapeBuffer offset:0 atIndex: 1];
|
||||
[encoder setBuffer: backend->tileCounters offset:0 atIndex: 2];
|
||||
[encoder setBuffer: backend->tilesArray offset:0 atIndex: 3];
|
||||
[encoder setBuffer: backend->triangleArray offset:0 atIndex: 4];
|
||||
[encoder setBuffer: backend->boxArray offset:0 atIndex: 5];
|
||||
[encoder setBytes: &clearColorVec4 length: sizeof(vector_float4) atIndex: 6];
|
||||
|
||||
//TODO: check that we don't exceed maxTotalThreadsPerThreadgroup
|
||||
DEBUG_ASSERT(RENDERER_TILE_SIZE*RENDERER_TILE_SIZE <= backend->computePipeline.maxTotalThreadsPerThreadgroup);
|
||||
|
|
@ -221,23 +239,27 @@ void mg_metal_canvas_viewport(mg_canvas_backend* interface, mp_rect viewPort)
|
|||
void mg_metal_canvas_update_vertex_layout(mg_metal_canvas_backend* backend)
|
||||
{
|
||||
char* vertexBase = (char*)[backend->vertexBuffer contents];
|
||||
char* shapeBase = (char*)[backend->shapeBuffer contents];
|
||||
char* indexBase = (char*)[backend->indexBuffer contents];
|
||||
|
||||
backend->interface.vertexLayout = (mg_vertex_layout){
|
||||
.maxVertexCount = MG_METAL_CANVAS_DEFAULT_BUFFER_LENGTH,
|
||||
.maxIndexCount = MG_METAL_CANVAS_DEFAULT_BUFFER_LENGTH,
|
||||
.posBuffer = vertexBase + offsetof(mg_vertex, pos),
|
||||
.posStride = sizeof(mg_vertex),
|
||||
.cubicBuffer = vertexBase + offsetof(mg_vertex, cubic),
|
||||
.cubicStride = sizeof(mg_vertex),
|
||||
.uvBuffer = vertexBase + offsetof(mg_vertex, uv),
|
||||
.uvStride = sizeof(mg_vertex),
|
||||
.colorBuffer = vertexBase + offsetof(mg_vertex, color),
|
||||
.colorStride = sizeof(mg_vertex),
|
||||
.posBuffer = vertexBase + offsetof(mg_vertex, pos),
|
||||
.posStride = sizeof(mg_vertex),
|
||||
.zIndexBuffer = vertexBase + offsetof(mg_vertex, zIndex),
|
||||
.zIndexStride = sizeof(mg_vertex),
|
||||
.clipBuffer = vertexBase + offsetof(mg_vertex, clip),
|
||||
.clipStride = sizeof(mg_vertex),
|
||||
.indexBuffer = [backend->indexBuffer contents],
|
||||
|
||||
.colorBuffer = shapeBase + offsetof(mg_shape, color),
|
||||
.colorStride = sizeof(mg_shape),
|
||||
.clipBuffer = shapeBase + offsetof(mg_shape, clip),
|
||||
.clipStride = sizeof(mg_shape),
|
||||
.uvBuffer = shapeBase + offsetof(mg_shape, uv),
|
||||
.uvStride = sizeof(mg_shape),
|
||||
|
||||
.indexBuffer = indexBase,
|
||||
.indexStride = sizeof(int)};
|
||||
}
|
||||
|
||||
|
|
@ -283,7 +305,10 @@ mg_canvas_backend* mg_metal_canvas_create(mg_surface surface)
|
|||
|
||||
//NOTE(martin): setup interface functions
|
||||
backend->interface.destroy = mg_metal_canvas_destroy;
|
||||
backend->interface.drawBuffers = mg_metal_canvas_draw_buffers;
|
||||
backend->interface.begin = mg_metal_canvas_begin;
|
||||
backend->interface.end = mg_metal_canvas_end;
|
||||
backend->interface.clear = mg_metal_canvas_clear;
|
||||
backend->interface.drawBatch = mg_metal_canvas_draw_batch;
|
||||
backend->interface.atlasUpload = mg_metal_canvas_atlas_upload;
|
||||
|
||||
mp_rect frame = mg_surface_get_frame(surface);
|
||||
|
|
@ -333,6 +358,9 @@ mg_canvas_backend* mg_metal_canvas_create(mg_surface surface)
|
|||
backend->vertexBuffer = [metalSurface->device newBufferWithLength: MG_METAL_CANVAS_DEFAULT_BUFFER_LENGTH*sizeof(mg_vertex)
|
||||
options: bufferOptions];
|
||||
|
||||
backend->shapeBuffer = [metalSurface->device newBufferWithLength: MG_METAL_CANVAS_DEFAULT_BUFFER_LENGTH*sizeof(mg_shape)
|
||||
options: bufferOptions];
|
||||
|
||||
backend->tilesArray = [metalSurface->device newBufferWithLength: RENDERER_TILE_BUFFER_SIZE*sizeof(int)*RENDERER_MAX_TILES
|
||||
options: MTLResourceStorageModePrivate];
|
||||
|
||||
|
|
|
|||
|
|
@ -20,14 +20,18 @@
|
|||
|
||||
typedef struct mg_vertex
|
||||
{
|
||||
vector_float2 pos; // position
|
||||
vector_float4 cubic; // canonical implicit curve space coordinates
|
||||
vector_float2 uv; // texture coordinates
|
||||
vector_float2 pos; // position
|
||||
int zIndex;
|
||||
} mg_vertex;
|
||||
|
||||
typedef struct mg_shape
|
||||
{
|
||||
vector_float4 color;
|
||||
vector_float4 clip;
|
||||
int zIndex;
|
||||
vector_float2 uv; // texture coordinates
|
||||
|
||||
} mg_vertex;
|
||||
} mg_shape;
|
||||
|
||||
typedef struct mg_triangle_data
|
||||
{
|
||||
|
|
|
|||
|
|
@ -35,9 +35,10 @@ bool is_top_left(float2 a, float2 b)
|
|||
|
||||
kernel void BoundingBoxKernel(constant mg_vertex* vertexBuffer [[buffer(0)]],
|
||||
constant uint* indexBuffer [[buffer(1)]],
|
||||
device mg_triangle_data* triangleArray [[buffer(2)]],
|
||||
device float4* boxArray [[buffer(3)]],
|
||||
constant float* contentsScaling [[buffer(4)]],
|
||||
constant mg_shape* shapeBuffer [[buffer(2)]],
|
||||
device mg_triangle_data* triangleArray [[buffer(3)]],
|
||||
device float4* boxArray [[buffer(4)]],
|
||||
constant float* contentsScaling [[buffer(5)]],
|
||||
uint gid [[thread_position_in_grid]])
|
||||
{
|
||||
uint triangleIndex = gid;
|
||||
|
|
@ -56,7 +57,9 @@ kernel void BoundingBoxKernel(constant mg_vertex* vertexBuffer [[buffer(0)]],
|
|||
float2 boxMax = max(max(p0, p1), p2);
|
||||
|
||||
//NOTE(martin): clip bounding box against clip rect
|
||||
vector_float4 clip = contentsScaling[0]*vertexBuffer[indexBuffer[vertexIndex]].clip;
|
||||
int shapeIndex = vertexBuffer[i0].zIndex;
|
||||
|
||||
vector_float4 clip = contentsScaling[0]*shapeBuffer[shapeIndex].clip;
|
||||
float2 clipMin(clip.x, clip.y);
|
||||
float2 clipMax(clip.x + clip.z-1, clip.y + clip.w-1);
|
||||
|
||||
|
|
@ -83,7 +86,7 @@ kernel void BoundingBoxKernel(constant mg_vertex* vertexBuffer [[buffer(0)]],
|
|||
//NOTE(martin): fill triangle data
|
||||
boxArray[triangleIndex] = float4(boxMin.x, boxMin.y, boxMax.x, boxMax.y);
|
||||
|
||||
triangleArray[triangleIndex].zIndex = vertexBuffer[i0].zIndex;
|
||||
triangleArray[triangleIndex].zIndex = shapeIndex;
|
||||
triangleArray[triangleIndex].i0 = i0;
|
||||
triangleArray[triangleIndex].i1 = i1;
|
||||
triangleArray[triangleIndex].i2 = i2;
|
||||
|
|
@ -174,11 +177,12 @@ float orient2d(float2 a, float2 b, float2 c)
|
|||
kernel void RenderKernel(texture2d<float, access::write> outTexture [[texture(0)]],
|
||||
texture2d<float> texAtlas [[texture(1)]],
|
||||
const device mg_vertex* vertexBuffer [[buffer(0)]],
|
||||
device uint* tileCounters [[buffer(1)]],
|
||||
const device uint* tilesArray [[buffer(2)]],
|
||||
const device mg_triangle_data* triangleArray [[buffer(3)]],
|
||||
const device float4* boxArray [[buffer(4)]],
|
||||
constant vector_float4* clearColor [[buffer(5)]],
|
||||
const device mg_shape* shapeBuffer [[buffer(1)]],
|
||||
device uint* tileCounters [[buffer(2)]],
|
||||
const device uint* tilesArray [[buffer(3)]],
|
||||
const device mg_triangle_data* triangleArray [[buffer(4)]],
|
||||
const device float4* boxArray [[buffer(5)]],
|
||||
constant vector_float4* clearColor [[buffer(6)]],
|
||||
uint2 gid [[thread_position_in_grid]],
|
||||
uint2 tgid [[threadgroup_position_in_grid]],
|
||||
uint2 threadsPerThreadgroup [[threads_per_threadgroup]],
|
||||
|
|
@ -274,12 +278,16 @@ kernel void RenderKernel(texture2d<float, access::write> outTexture [[texture(0)
|
|||
float4 cubic1 = v1->cubic;
|
||||
float4 cubic2 = v2->cubic;
|
||||
|
||||
float2 uv0 = v0->uv;
|
||||
float2 uv1 = v1->uv;
|
||||
float2 uv2 = v2->uv;
|
||||
|
||||
int zIndex = v0->zIndex;
|
||||
float4 color = v0->color;
|
||||
float4 color = shapeBuffer[zIndex].color;
|
||||
|
||||
/////////////////////////////////////////////////////////////////////////
|
||||
//TODO: dummy uv while we figure out image handling.
|
||||
float2 uv0 = shapeBuffer[zIndex].uv;
|
||||
float2 uv1 = uv0;
|
||||
float2 uv2 = uv0;
|
||||
/////////////////////////////////////////////////////////////////////////
|
||||
|
||||
|
||||
for(int i=0; i<6; i++)
|
||||
{
|
||||
|
|
|
|||
|
|
@ -9,6 +9,8 @@
|
|||
#ifndef __METAL_SURFACE_H_
|
||||
#define __METAL_SURFACE_H_
|
||||
|
||||
#include"graphics.h"
|
||||
|
||||
#ifdef __OBJC__
|
||||
#import<Metal/Metal.h>
|
||||
#endif
|
||||
|
|
|
|||
Loading…
Reference in New Issue