[번역]Mesh Shaders and Meshlet Culling in Metal 3

역자 주.
번역글에 들어 가기전에... 먼저 그것이 무엇인가요? 에 대한 간단한 개요 또는 설명서?가 있어야겠다 싶습니다.

meshlet culling 은 컴퓨터 그래픽에서 그래픽 하드웨어가 처리하고 렌더링해야 하는 트라이앵글의 수를 줄여 렌더링 성능을 향상시키는 데 사용되는 기술입니다. 특히 효율성이 중요한 비디오 게임과 같은 실시간 애플리케이션에서 유용합니다.

기존 렌더링 파이프라인에서는 3D 씬의 모든 트라이앵글을 렌더링하는 경우가 많기 때문에 계산 비용이 많이 들 수 있습니다. meshlet 컬링은 삼각형을 'Meshlet'이라는 작은 단위로 그룹화한 다음 보이지 않거나 최종 이미지에 거의 기여하지 않는 전체 meshlet 을 컬링하거나 건너뛰는 방식으로 이 프로세스를 최적화하는 것을 목표로 합니다.

meshlet  컬링의 일반적인 작동 방식은 다음과 같습니다:

meshlet  생성: 첫 번째 단계는 3D 메시를 meshlet 으로 분할하는 것입니다. meshlet 은 일반적으로 적은 수의 트라이앵글(예: 64개 이하)을 포함하도록 설계되며 함께 렌더링될 가능성이 있는 트라이앵글을 공간적으로 그룹화하는 방식으로 구성됩니다.

컬링: 메쉬렛이 생성되면 컬링 알고리즘을 적용하여 특정 메쉬렛을 렌더링할지 여부를 결정합니다. 뷰 프러스텀 컬링(메시가 카메라의 시야 밖에 있는지 확인), 오클루전 컬링(메시가 다른 오브젝트에 가려져 있는지 확인), LOD(세부 수준) 컬링(거리에 따라 메시의 세부 수준이 낮은 버전 선택) 등 다양한 컬링 기법을 사용할 수 있습니다.

렌더링: 컬링 후에는 가시적이고 관련성 있는 meshlet 만 그래픽 하드웨어로 전송하여 렌더링하므로 씬의 모든 트라이앵글을 렌더링하는 것에 비해 작업량이 크게 줄어듭니다.

meshlet  컬링은 시각적 품질을 유지하면서 렌더링 성능을 최적화하는 데 도움이 될 수 있습니다. meshlet 컬링은 최신 그래픽 엔진에서 효율적이고 시각적으로 매력적인 실시간 렌더링을 구현하기 위해 하드웨어 가속 레이 트레이싱, LOD, 동적 레벨 오브 디테일 조정과 같은 다른 기술과 함께 사용되는 경우가 많습니다.

meshlet 컬링은 특정 기술 및 개념으로, 단일한 발명가나 제작자가 존재하지 않습니다. 컴퓨터 그래픽 분야에서 연구자와 개발자가 비디오 게임과 같은 실시간 애플리케이션의 렌더링 성능을 개선할 방법을 모색하면서 시간이 지남에 따라 발전해 온 그래픽 최적화 기법입니다.

---

원문.

Mesh Shaders and Meshlet Culling in Metal 3

Apple이 WWDC 2022에서 소개한 Metal 3에는 최신 렌더링 기법, 더 빠른 리소스 로딩, 유연한 셰이더 컴파일을 가능하게 하는 다양한 기능이 포함되어 있습니다. 또한 개발자가 대부분의 기존 버텍스 처리 단계를 건너뛰고 지오메트리를 래스터라이저에 직접 제출할 수 있는 완전히 새로운 지오메트리 파이프라인이 포함되어 새로운 렌더링 기법을 활용할 수 있습니다.

Discover Metal 3 - WWDC22 - Videos - Apple Developer

이 글에서는 Metal 3의 새로운 지오메트리 파이프라인의 기능과 작동 방식, 사용 사례에 대해 살펴보겠습니다. 그런 다음 메시 셰이더를 사용하여 최신 GPU 기반 렌더링 엔진의 중요한 기능인 meshlet 컬링을 구현하는 방법에 대해 자세히 살펴보겠습니다.

이 글의 샘플 코드는 여기에서 다운로드하세요. 간결성을 위해 일부 구현 세부 사항은 이 설명에서 생략했으며, 코드를 읽어보시는 것을 대신할 수는 없습니다.

새로운 지오메트리 파이프라인 Metal의 기본 렌더링 기능 사용에 익숙하다면 버텍스 디스크립터를 사용해 보셨을 것입니다. 
버텍스 디스크립터는 버텍스 어트리뷰트가 버텍스 버퍼에 배치되는 방식을 나타냅니다. 

렌더 파이프라인 디스크립터에 버텍스 디스크립터를 포함하면 셰이더 컴파일러가 현재 버텍스의 데이터를 버텍스 함수의 스테이지인 인수에 자동으로 로드하는 코드(버텍스 함수 프리앰블)를 삽입할 수 있습니다. 

이 기능을 버텍스 페치라고 합니다. 반대로 버텍스 설명자를 사용하지 않을 때는 버텍스 버퍼에서 버텍스 데이터를 직접 수동으로 로드해야 하는데, 이를 버텍스 풀이라고 합니다.

Vertex Formats Part 2: Fetch vs Pull – Yosoygames

버텍스 가져오기를 사용하면 버텍스 함수를 간단하게 작성할 수 있습니다. 이 기능이 없으면 데이터를 수동으로 로드해야 할 뿐만 아니라 필요한 변환(예: 정규화된 정수 표현에서 부동 소수점으로 또는 3요소 벡터에서 4요소 벡터로)도 수동으로 수행해야 합니다. 하지만 이 접근 방식에도 단점이 있습니다.

동기 기존 프로그래머블 그래픽스 파이프라인의 버텍스 처리 부분은 많은 종류의 씬에 최적이 아닙니다. 

드로우 콜에는 인코딩 오버헤드가 발생하기 때문에 역사적으로 드로우 콜 수를 줄이는 것이 권장되어 왔습니다. 

이는 곧 각 드로우 콜이 넓은 영역에 걸쳐 있고 임의의 방향으로 향하는 삼각형을 포함하는 지오메트리를 그릴 수 있다는 것을 의미합니다. 

버텍스는 프리미티브로 조립되기 전에 처리되어야 하므로 즉시 컬링될 삼각형에 속한 버텍스 데이터를 로드하는 데 메모리 대역폭이 낭비됩니다. 

또한 버텍스 함수는 버텍스 수준에서 작동하기 때문에 프리미티브에 대해 추론할 방법이 없고 중간에 프리미티브를 효율적으로 거부할 방법이 없습니다.

컬링에서 살아남은 삼각형의 경우에도 전통적인 접근 방식은 변환 후 정점 캐시를 과소 활용하는 경우가 많습니다.

인덱스 버퍼는 종종 정렬되지 않기 때문에 인접한 인덱스는 인접한 프리미티브에 속하는 정점을 참조합니다.

이는 정점 캐시의 관점에서 Triangle soup 처럼 보이는 인덱스 메쉬로 이어집니다.

Triangle soup

2015년부터 지오메트리를 래스터라이저로 전달하기 전에 트라이앵글 가시성을 결정하기 위해 컴퓨팅 셰이더를 사용하는 것이 점점 더 강조되고 있습니다.

영향력 있는 GDC 2016 강연자인 그레이엄 윌리달의 강연은 AMD의 GCN 아키텍처에 대한 기술적이고 구체적인 내용이지만 컴퓨팅 기반 지오메트리 처리의 사고방식을 익히는 데 유용한 참고 자료가 됩니다.

아래에서 meshlet 컬링에 대해 설명하면서 이러한 아이디어 중 몇 가지를 Metal에서 구현해 보겠습니다.

컴퓨팅 기반 지오메트리 처리 기술이 보편화되면서 GPU 공급업체와 그래픽 API 디자이너는 이러한 아이디어를 제품에 통합하기 시작했고, 최신 지오메트리 파이프라인이 탄생했습니다. 

Apple의 컴퓨팅 기반 지오메트리 처리 구현은 그래픽 파이프라인에 래스터화와 함께 인라인(동일한 명령 인코더에서)으로 실행할 수 있는 프로그래밍 가능한 두 단계를 추가합니다. 

이러한 단계는 오브젝트 셰이더와 메시 셰이더이며, 각 단계를 차례로 살펴보겠습니다.

역자 주.
세 줄 요약.
1. Metal은 고성능 그래픽스 렌더링을 위한 API로, 지오메트리 파이프라인의 최적화 기술을 제공합니다.
2. 이 기술은 버텍스 데이터의 효율적인 처리와 래스터라이제이션 이전에 트라이앵글 가시성을 결정하는 데 사용됩니다.
3. 컴퓨팅 기반 지오메트리 처리를 통해 렌더링 성능을 향상시킬 수 있으며, Metal은 이를 지원합니다.

오브젝트 함수 "메쉬"라는 단어는 다소 모호합니다. 흔히 메시를 3D 모델로 생각하는데, 메시에는 여러 개의 서브메시가 포함될 수 있으며 각 서브메시에는 고유한 머티리얼 속성이 있는 경우가 많습니다.

오브젝트 및 메시 셰이더의 맥락에서 "메시"라는 단어는 좀 더 특별한 의미를 갖습니다. 

메쉬는 정점, 인덱스 및 프리미티브별 데이터의 작은 모음입니다. 메쉬는 풀 한 포기, 머리카락 한 가닥 또는 최대 수십 개의 정점과 수십 개의 삼각형으로 구성된 기타 모양일 수 있습니다. 

메시의 크기가 제한되는 이유는 효율적인 처리를 위해 스레드 그룹 메모리에 맞출 수 있어야 하기 때문입니다.

오브젝트 셰이더의 주요 아이디어는 프로그래머가 나머지 파이프라인을 통해 어떤 지오메트리 청크를 진행할지 결정할 수 있는 기회를 제공하는 것입니다. 

컴퓨팅 그리드와 마찬가지로 오브젝트는 애플리케이션에서 의미 있는 모든 작업 모음을 나타낼 수 있는 추상 엔티티입니다. 

오브젝트는 모델 모음, 

더 큰 모델의 단일 서브메시, 

지형 패치 또는 하나 이상의 메시를 생성할 수 있는 기타 작업 단위일 수 있습니다.

오브젝트 셰이더의 핵심 개념은 증폭이라는 개념입니다. 

오브젝트 셰이더 실행 모델 때문에 각 오브젝트는 0개, 1개 또는 다수의 메시를 스폰할 수 있습니다.

각 버텍스 또는 프래그먼트가 버텍스 또는 프래그먼트 함수의 호출로 처리되는 것처럼 오브젝트는 오브젝트 함수의 호출로 처리됩니다. 

스레드 그룹 크기와 스레드 그룹 수 및 계산 파이프라인을 지정하여 계산 그리드를 디스패치하는 방식과 유사하게, 렌더링 명령 인코더에서 실행할 오브젝트 스레드 그룹의 크기와 실행할 오브젝트 스레드 그룹 수를 취하는 메서드를 호출하여 오브젝트를 그립니다.

각 오브젝트 스레드는 추가 처리를 위해 메시 함수 호출에 전달되는 임의의 데이터 집합인 페이로드를 생성합니다. 

아래에서 오브젝트 셰이더를 작성하는 방법을 살펴보겠지만, 요점은 오브젝트 셰이더는 오브젝트가 생성해야 하는 메시 수를 결정하고 이후에 실행되는 각 메시 스레드에 페이로드 데이터를 제공함으로써 증폭을 달성한다는 점입니다.

메시 함수 메시 함수는 개별 정점 대신 정점 그룹에서 작동할 수 있는 새로운 유형의 셰이더 함수입니다. 

위에서 언급한 바와 같이 메시란 메시 셰이더에서 생성되어 고정 함수 래스터라이저로 전달되는 정점, 인덱스 및 프리미티브별 데이터의 집합을 말합니다.

오브젝트 함수가 0개 이상의 메시를 설명하는 페이로드를 생성할 수 있는 것처럼 메시 함수는 0개 이상의 프리미티브(구성 버텍스에서 함께 스티칭된)로 구성된 메시를 생성할 수 있습니다. 

객체 함수에 의해 유발된 각 메쉬 스레드 그룹은 단일(잠재적으로 비어 있는) 메쉬를 생성합니다.

메시 스레드 그룹의 스레드는 집합적으로 다음과 같은 작업을 수행합니다:

- 버텍스 데이터를 출력 메시로 복사 
- 인덱스 데이터를 출력 메시로 복사 
- 프리미티브별 데이터를 출력 메시로 복사 
- 메시가 포함하는 총 프리미티브 수 설정

메시 생성 작업을 스레드 간에 분할하는 방법은 사용자에게 달려 있습니다.

특히 작은 메시의 경우 단일 스레드가 모든 작업을 수행할 수 있습니다. 각 스레드가 버텍스, 몇 개의 인덱스 및/또는 프리미티브를 생성하여 스레드 간에 부하를 공유하는 것이 더 효율적일 때가 많습니다.

아래 meshlet 컬링 섹션에서 작업을 분할하는 방법에 대한 자세한 예를 살펴보겠습니다.

메시 셰이더 사용 사례 메시 셰이더는 개별 트라이앵글보다 더 거친 수준에서 지오메트리를 처리할 때 유용합니다.

메시 셰이더는 헤어, 퍼, 폴리지 또는 파티클 트레이스와 같은 절차적 지오메트리를 생성할 수 있습니다.

화면 공간 범위 및 카메라와의 거리와 같은 메트릭을 기반으로 미리 계산된 디테일 레벨 중에서 선택하는 데 사용할 수 있습니다.

또한 공간 일관성을 활용하여 버텍스 재사용을 완전히 활용하고 뒷면을 향한 트라이앵글을 처리하는 데 낭비되는 작업을 피하는 meshlet 을 생성할 수 있습니다.

이러한 각 사용 사례를 차례로 살펴보겠습니다.

프로시저럴 지오메트리 
메시 셰이더의 가장 중요한 사용 사례 중 하나는 프로시저럴 지오메트리입니다. 프로시저럴 지오메트리는 미리 만들어진 에셋을 사용하지 않고 알고리즘적으로 지오메트리를 생성하는 프로세스 및 기법의 범주입니다. 

메시 셰이더는 셰이프의 전체 표현을 메모리에 보관하는 대신 즉시 셰이프를 생성하여 씬의 메모리 공간을 늘리지 않고도 씬의 디테일을 크게 향상시킬 수 있습니다. 프로시저럴 메서드는 그래픽 분야에서 오랫동안 사용되어 왔으며, 이제 완전히 확장된 지오메트리를 메모리에 보관하지 않고도 단순화된 표현에서 세부적인 지오메트리를 생성할 수 있기 때문에 메시 셰이더와 함께 그 매력이 더욱 커졌습니다. 

퍼 렌더링의 맥락에서 프로시저럴 지오메트리의 예는 메시 셰이딩에 대한 WWDC 2022 세션을 참조하세요.

Transform your geometry with Metal mesh shaders - WWDC22 - Videos - Apple Developer

LOD 선택 메시 셰이더의 또 다른 사용 사례는 세부 수준(LOD) 선택입니다. LOD 선택은 카메라로부터의 거리 또는 기타 측정값에 따라 오브젝트에 적합한 디테일 레벨을 선택하는 프로세스입니다. 디테일 레벨은 밉맵과 유사한 이산 레벨 세트에서 미리 계산하거나 고정 함수 테셀레이션과 유사한 파라메트릭 표현에서 즉석에서 생성할 수 있습니다.

위에서 언급한 메시 셰이더에 대한 WWDC 세션에 포함된 샘플 코드는 세부 수준 선택을 구현하는 방법에 대한 초보적인 예시를 제공합니다.

이제 메시 셰이더의 몇 가지 사용 사례(프로시저럴 지오메트리와 세부 수준 선택)에 대해 언급했으니 이제 점점 더 보편적으로 사용되는 새로운 지오메트리 파이프라인인 meshlet 컬링에 대해 자세히 알아보겠습니다.

메시클릿과 메시클릿 컬링 메시 셰이더의 가장 흥미로운 사용 사례 중 하나는 메시클릿 컬링입니다. 이 기술의 장점을 이해하려면 먼저 meshlet 이 무엇인지 알아야 합니다.

대부분의 3D 모델링 패키지는 몇 가지 일반적인 형식(웨이브프론트 .obj, glTF, USD[Z] 등)의 에셋을 생성합니다. 이러한 프로그램에서 생성되는 메시의 대부분은 정돈되지 않은 정점과 인덱스의 목록으로, 프로그램의 내부 표현을 반영하는 방식으로(최적의 순서가 아닌) 서로 연결되는 경우가 많습니다. 

이러한 에셋을 로드하고 렌더링하면 특정 메시의 많은 트라이앵글이 카메라를 향하고 있을 가능성이 높으며, 인덱스 버퍼가 버텍스 캐시를 최적으로 사용하는 순서로 버텍스를 참조하지 못할 가능성이 높습니다.

어떻게 해야 할까요? 최근 몇 년 동안 메시를 meshlet 으로 세분화하는 것이 더 보편화되었습니다. 이름에서 알 수 있듯이 meshlet 은 더 큰 메시를 집합적으로 구성하는 작은 메시입니다. 중요한 점은 메쉬렛은 일관성을 위해 구성된다는 것입니다. 메시의 정점은 서로 가까이 있어야 하고, 메시의 인덱스는 정점의 인접성과 일치하도록 배치되어야 하며, 메시 삼각형의 법선은 서로 같은 일반적인 방향을 가리켜야 합니다.

메시를 meshlet 으로 전처리 메시를 meshlet 으로 바꾸려면 어떻게 해야 할까요? 기본적으로 공간적으로 일관된 정점을 순서대로 참조할 수 있도록 메시의 정점을 재정렬하고, 정점을 삼각형으로 연결하는 작은 인덱스 버퍼를 구축하는 것으로 요약할 수 있습니다. 따라서 각 메슬렛은 해당 정점을 포함하는 원본 정점 버퍼의 스팬에 대한 참조와 해당 삼각형을 구성하는 인덱스 트리플 목록으로 구성됩니다. 메슬렛은 제한된 수의 정점만 참조할 수 있으므로 이러한 인덱스는 일반적으로 인덱스 드로잉을 수행할 때 일반적으로 사용하는 16~32비트 인덱스보다 작습니다. 샘플 코드에서는 8비트 부호 없는 정수를 인덱스로 사용합니다.

오픈 소스 메쉬 옵티마이저 라이브러리는 메쉬를 메쉬렛으로 분할하는 유연하고 효율적인 도구입니다. 여기서는 메쉬옵티마이저의 모든 기능을 자세히 살펴보지는 않고 가장 간단한 메쉬렛 생성 함수인 meshopt_buildMeshlets를 사용하겠습니다. 이 함수는 인덱싱된 메쉬 또는 서브메쉬를 받아 다음을 생성합니다:

- 최적의 순서로 정렬된 정점을 원래 정점 목록의 위치에 매핑하는 meshlet 정점 목록, 
- 각 삼각형에 대해 3개씩 8비트 인덱스 목록인 meshlet 삼각형 목록, 
- 각각 정점의 범위와 삼각형의 범위를 참조하는 meshlet 목록, 
- 각 meshlet 의 대략적인 경계 구, 
- 각 meshlet 정점 법선의 평균 방향과 퍼짐을 나타내는 원뿔입니다.

이러한 출력은 메탈 버퍼에 복사하여 오브젝트 및 메시 셰이더에서 직접 사용할 수 있습니다. 아래 그림은 meshlet 버텍스 버퍼가 인덱스를 원래 버텍스 버퍼에 매핑하는 방법과 meshlet 이 삼각형 목록과 버텍스 목록의 각 부분을 나타내는 방법을 보여줍니다. 한 가지 흥미로운 점은 메슬렛 정점(위쪽 화살표 집합)에서의 매핑은 다소 일관성이 없지만(메모리에 흩어져 있음), 메슬렛 내의 인덱스에 의한 참조는 매우 일관성 있고 밀도가 높다는 점입니다.

meshlet 이 정확히 어떻게 생성되는지 궁금하다면 샘플 코드의 meshlet 생성 대상을 살펴보세요2. 모델 I/O를 사용하여 3D 모델을 로드하고 사전 처리된 "메시화된" 메시를 사용자 지정 형식으로 생성하는 작은 명령줄 유틸리티입니다.

meshlet  컬링 기법 메시를 meshlet 으로 잘게 쪼개고 나면 이를 사용하여 렌더링을 더 효율적으로 만들려면 어떻게 해야 할까요? 오브젝트 셰이더와 함께 메시의 공간 일관성과 노멀 일관성을 활용하여 보이지 않는 메시의 정점을 처리하는 데 시간을 소비하기 전에 메시를 컬링합니다. 이를 위해 프러스텀 컬링과 노멀 콘 컬링을 수행합니다.

meshlet  프러스텀 컬링은 뷰 볼륨(즉, 뷰 프러스텀)을 meshlet 의 바운딩 구를 저렴하게 테스트할 수 있는 평면 집합으로 변환하는 방식으로 수행됩니다. 바운딩 구체가 프러스텀 평면의 음의 절반 공간에 완전히 놓여 있으면 뷰 볼륨에 포함되지 않으므로 컬링할 수 있습니다.

일반 원뿔 컬링은 약간 더 복잡합니다.
meshlet 전처리 단계의 일부로 각 meshlet 에 대해 삼각형의 평균 법선 방향을 따라 방향이 지정된 원뿔을 생성합니다. 원뿔의 너비는 평균 법선과 정점 법선 사이의 최대 확산을 나타냅니다. 이 정보를 사용하면 노멀 원뿔이 카메라에서 충분히 멀리 떨어져 있는 메시를 컬링할 수 있습니다.
노멀 원뿔에 카메라가 포함되지 않으면 정의상 카메라가 메시의 어떤 면도 볼 수 없습니다. 이는 메시 집합의 모든 삼각형을 한 번에 고려하는 집계 백페이스 컬링의 한 형태입니다. 제가 알기로는 1993년 시르문과 아비 에지가 베지어 패치 컬링의 맥락에서 그래픽에 도입한 것입니다.

오브젝트 및 메시 셰이더의 기본 사항에 대해 간단히 소개한 후 아래에서 오브젝트 셰이더에서 이 두 가지 컬링 기법을 구현하는 방법을 살펴보겠습니다.

메시 렌더 파이프라인 상태 생성 메시 셰이더를 통합하는 렌더 파이프라인 상태를 생성하는 것은 기존 지오메트리 파이프라인을 사용하여 파이프라인 상태를 생성하는 것과 매우 유사합니다. 가장 큰 차이점은 버퍼에서 버텍스 데이터를 수동으로 로드하기 때문에 vertex discriptor 를 포함하지 않는다는 점입니다.

어태치먼트 픽셀 형식과 블렌딩 상태를 설정하는 일반적인 작업 외에도 오브젝트, 메시, 프래그먼트 함수를 생성하고 렌더 파이프라인 디스크립터의 해당 프로퍼티에 이를 설정합니다:

id<MTLFunction> objectFunction = [library newFunctionWithName:@"my_object_function"];
id<MTLFunction> meshFunction = [library newFunctionWithName:@"my_mesh_function"];
id<MTLFunction> fragmentFunction = [library newFunctionWithName:@"my_fragment_function"];
 
MTLMeshRenderPipelineDescriptor *pipelineDescriptor = [MTLMeshRenderPipelineDescriptor new];
 
pipelineDescriptor.objectFunction = objectFunction;
pipelineDescriptor.meshFunction = meshFunction;
pipelineDescriptor.fragmentFunction = fragmentFunction;

그런 다음 디바이스에서 새로운 -newRenderPipelineStateWithMeshDescriptor: options:reflection:error: 메서드를 사용하여 메시 렌더 파이프라인 상태를 가져올 수 있습니다:

[device newRenderPipelineStateWithMeshDescriptor:pipelineDescriptor
                                         options:MTLPipelineOptionNone 
                                      reflection:nil
                                           error:&error]

오브젝트 및 메시 리소스 바인딩 오브젝트 및 메시 함수는 다른 종류의 셰이더 함수와 마찬가지로 리소스를 참조할 수 있습니다. 메시 셰이더는 이를 위해 MTLRenderCommandEncoder 프로토콜에 다음과 같은 몇 가지 새로운 메서드를 추가했습니다:

- (void)setObjectBytes:(const void *)bytes 
                length:(NSUInteger)length
               atIndex:(NSUInteger)index;
- (void)setObjectBuffer:(id <MTLBuffer>)buffer 
                 offset:(NSUInteger)offset 
                atIndex:(NSUInteger)index;
- (void)setObjectTexture:(id <MTLTexture>)texture 
                atIndex:(NSUInteger)index;
- (void)setMeshBytes:(const void *)bytes 
              length:(NSUInteger)length 
             atIndex:(NSUInteger)index;
- (void)setMeshBuffer:(id <MTLBuffer>)buffer 
               offset:(NSUInteger)offset 
              atIndex:(NSUInteger)index
- (void)setMeshTexture:(id <MTLTexture>)texture 
               atIndex:(NSUInteger)index

이러한 리소스를 바인딩하는 것은 다른 프로그래밍 가능한 단계와 동일하게 작동합니다.

메시 드로 콜 2계층 오브젝트/메시 실행 모델이 전체 기능의 핵심이기 때문에 메시 드로 콜의 구조를 이해하는 것이 중요합니다.

메탈 메시 셰이더는 MTLRenderCommandEncoder 프로토콜에 몇 가지 새로운 드로 메서드를 추가합니다. 여기서는 그 중 하나인 -drawMeshThreadgroups:스레드당 오브젝트 스레드 그룹: 스레드당 메시 스레드 그룹만 사용하겠습니다. 시그니처는 다음과 같습니다:

-(void)drawMeshThreadgroups:(MTLSize)threadgroupsPerGrid 
threadsPerObjectThreadgroup:(MTLSize)threadsPerObjectThreadgroup
  threadsPerMeshThreadgroup:(MTLSize)threadsPerMeshThreadgroup;

스레드 그룹별 그리드 파라미터는 메탈에 얼마나 많은 오브젝트 스레드 그룹을 시작해야 하는지 알려줍니다. 각 오브젝트 스레드 그룹은 궁극적으로 0개, 1개 또는 다수의 메시 스레드 그룹을 시작할 수 있다는 점을 기억하세요.

스레드당 오브젝트 스레드 그룹 매개변수는 각 오브젝트 스레드 그룹에 있는 스레드 수를 지정합니다. 컴퓨팅 커널과 마찬가지로 이 숫자는 렌더 파이프라인 상태의 objectThreadExecutionWidth 프로퍼티에서 검색할 수 있는 파이프라인 스레드 실행 폭의 배수여야 합니다(현재 Apple GPU의 경우 일반적으로 32입니다).

스레드당 메시 스레드 그룹 파라미터는 각 메시 스레드 그룹에 있는 스레드 수를 지정합니다. 이전 파라미터와 마찬가지로 스레드 실행 폭의 배수여야 하며, 이 경우 MTLRenderPipelineState의 meshThreadExecutionWidth 프로퍼티로 사용할 수 있습니다.

이 드로 콜에 의해 실행될 메시 스레드 그룹의 수는 지정하지 않았습니다. 결국 오브젝트 셰이더의 핵심은 오브젝트 함수 자체가 처리할 메시 수를 결정한다는 점입니다.

오브젝트 함수 오브젝트 함수의 핵심 구조는 다음과 같습니다:

[[object]]
void my_object_function(
    object_data Payload &object [[payload]],
    grid_properties grid)
{
    // Optionally populate the object's payload
    object.someProperty = ...;
    // Set the output grid's threadgroup count
    // (Only do this from one object thread!)
    grid.set_threadgroups_per_grid(uint3(meshCount, 1, 1));
}

객체 함수 앞에는 객체 함수로 표시하는 새로운 [[object]] 속성이 붙습니다.

객체 함수는 [[페이로드]] 속성을 가진 매개변수를 받을 수 있습니다. 이 매개변수가 있는 경우 이 매개변수는 object_data 주소 공간에 있는 참조 또는 포인터여야 합니다. 이 파라미터의 유형은 사용자가 제어하며, 메시 셰이더가 호출하는 오브젝트 함수에서 참조해야 할 수 있는 데이터를 포함하는 구조체입니다. 함수 본문에서 원하는 대로 이 파라미터를 채우면 됩니다. 아래에서는 이 파라미터를 사용하여 메시 셰이더에 렌더링할 meshlet 을 알려줍니다.

오브젝트 함수는 또한 단일 메서드인 set_threadgroups_per_grid가 있는 grid_properties 유형의 파라미터를 받습니다. 이것은 객체 함수가 그리드 스레드 그룹을 디스패치하도록 하는 메커니즘입니다. 스레드 그룹 수를 0이 아닌 값으로 설정하면 Metal이 파이프라인의 메시 함수의 스레드 그룹을 해당 수만큼 실행하도록 지시합니다.

중요한 점은 각 오브젝트 스레드 그룹에 하나의 스레드만 스레드 그룹 수를 채워야 한다는 것입니다. 현재 스레드의 위치를 확인하고 스레드의 위치가 0일 때만 이 프로퍼티를 작성할 수 있도록 [[thread_position_in_threadgroup]]으로 어트리뷰션된 파라미터를 오브젝트 함수에 추가할 수 있습니다(샘플 코드에서 이를 보여줌).

meshlet 컬링 오브젝트 셰이더 meshlet 컬링 오브젝트 함수가 하는 일은 컬링 기법에 대한 섹션에서 설명한 대로 메 meshlet 시렛 컬링을 수행하는 것입니다. 렌더링 중인 메시의 각 meshlet 에 대해 메시 함수가 처리할지 여부를 결정하기에 충분한 정보만 로드합니다.

아래 설명에서는 함수에 필요한 버퍼가 현재 렌더 명령 인코더에 적절히 바인딩되어 있다고 가정하므로 자세한 내용은 샘플 코드를 참조하세요.

meshlet 에 속한 모든 데이터를 캡슐화하는 다음과 같은 구조가 있다고 가정해 보겠습니다:

struct MeshletDescriptor {
    uint vertexOffset;
    uint vertexCount;
    uint triangleOffset;
    uint triangleCount;
    packed_float3 boundsCenter;
    float boundsRadius;
    packed_float3 coneApex;
    packed_float3 coneAxis;
    float coneCutoff;
    //...
};

오프셋과 카운트 멤버는 각각 메쉬렛 버텍스 버퍼와 메쉬렛 트라이앵글 버퍼 내의 스팬을 참조합니다. 이들은 오브젝트 셰이더에서 사용되지 않습니다. 바운딩 프로퍼티와 콘 프로퍼티만 사용하여 컬링을 수행할 것입니다.

오브젝트의 페이로드를 저장할 커스텀 타입도 필요합니다. 이는 컬링 테스트를 통과한 meshlet 인덱스 목록으로 간단히 구성됩니다:

struct ObjectPayload {
    uint meshletIndices[kMeshletsPerObject];
};

설명을 위해 객체 함수를 약간 단순화하겠습니다. 전체 구현은 샘플 코드를 참조하세요. 객체 함수 시그니처는 다음과 같습니다:

[[object]]
void object_main(
    device const MeshletDescriptor *meshlets [[buffer(0)]],
    constant InstanceData &instance          [[buffer(1)]],
    uint meshletIndex                        [[thread_position_in_grid]],
    uint threadIndex                         [[thread_position_in_threadgroup]],
    object_data ObjectPayload &outObject     [[payload]],
    mesh_grid_properties outGrid)

이전과 마찬가지로 페이로드 매개변수와 메시 그리드 속성 매개변수가 있음을 알 수 있습니다. 또한 meshlet 메타데이터가 포함된 버퍼와 일부 인스턴스별 데이터가 포함된 작은 버퍼에 대한 포인터를 가져옵니다.

객체 함수 본문에서는 객체 그리드에서 스레드의 위치를 사용하여 컬링을 수행할 meshlet 을 검색합니다:

device const MeshletDescriptor &meshlet = meshlets[meshletIndex];

여기서는 프러스텀 컬링과 노멀 콘 컬링의 세부 사항은 중요하지 않으며, 작은 유틸리티 함수를 호출하여 두 가지를 모두 수행합니다:

bool frustumCulled = !sphere_intersects_frustum(frustumPlanes, meshlet.boundsCenter, meshlet.boundsRadius);
 
bool normalConeCulled = cone_is_backfacing(meshlet.coneApex, meshlet.coneAxis, meshlet.coneCutoff, cameraPosition);

많은 메시에 대해 동시에 작업하고 있으므로 각 메시에 대해 페이로드 배열의 적절한 인덱스에 쓰도록 객체 스레드를 조정해야 합니다.

컬링 결과를 하나의 정수 값으로 결합하는 것부터 시작합니다:

int passed = (!frustumCulled && !normalConeCulled) ? 1 : 0;

그런 다음 접두사 합산 연산을 사용하여 우리보다 인덱스가 작은 스레드 중 컬링 테스트를 통과한 스레드 수를 확인합니다. 접두사 합산에 대해 잘 모르신다면 다음과 같은 리소스를 참조하세요.

int payloadIndex = simd_prefix_exclusive_sum(passed);

결과 페이로드 인덱스는 meshlet 이 컬링되지 않은 경우 스레드가 해당 meshlet 의 인덱스를 어디에 써야 하는지 알려줍니다. 따라서 전달된 값을 참조한 다음 쓰기를 수행합니다:

if (passed) {
    outObject.meshletIndices[payloadIndex] = meshletIndex;
}

객체 함수의 마지막 작업은 메시 스레드 그룹 수를 쓰는 것입니다. 위에서 언급했듯이 하나의 스레드만 이 작업을 수행해야 하므로 먼저 컬링되지 않은 메시의 총 수를 계산한 다음, 오브젝트 스레드 그룹의 첫 번째 스레드인 경우 메시 셰이더 호출 수를 기록합니다:

uint visibleMeshletCount = simd_sum(passed);
if (threadIndex == 0) {
    outGrid.set_threadgroups_per_grid(uint3(visibleMeshletCount, 1, 1));
}

이것으로 객체 함수의 본문이 끝났습니다. 이제 페이로드에는 렌더링할 메시의 인덱스가 포함되고 그리드 프로퍼티에는 실행할 메시 스레드 그룹 수가 포함됩니다.

메시 셰이더 출력 메시 함수에는 함수에 의해 생성된 버텍스, 인덱스 및 프리미티브별 데이터를 수집하는 사용자 정의 유형의 파라미터가 있습니다. 메시의 스레드 그룹의 스레드가 협업하여 이 데이터를 생성합니다. 정점 데이터와 프리미티브별 데이터를 집계하는 유형을 지정하여 메시의 유형을 정의합니다.

먼저 버텍스 데이터를 정의합니다. 이는 일반적인 버텍스 함수의 반환 유형과 비슷합니다.

struct MeshletVertex {
    float4 position [[position]];
    float3 normal;
    float2 texCoords;
};

간단한 예제에서는 시각화 목적으로 전달되는 유일한 프리미티브별 데이터는 색상입니다:

struct MeshletPrimitive {
    float4 color [[flat]];
};

이러한 구조는 metal::mesh 클래스의 템플릿 인스턴스화로 구성된 typedef를 선언하여 출력 메시로 통합됩니다:

using Meshlet = metal::mesh<MeshletVertex, MeshletPrimitive, kMaxVerticesPerMeshlet, kMaxTrianglesPerMeshlet, topology::triangle>;

 

Meshlet Mesh Shader
파이프라인의 이 단계에서 Metal은 각 오브젝트 스레드 그룹이 지정한 스레드 그룹 수가 포함된 메시 그리드를 시작합니다. 각 메시 스레드 그룹의 스레드 수는 그리기 호출을 인코딩할 때 지정되므로 이 숫자는 단일 meshlet 을 처리하는 데 필요한 최대 스레드 수여야 합니다3.

메시 스레드 그룹의 스레드 수는 메시의 최대 정점 및 삼각형 수와 메시 생성 작업이 메시 스레드에 분산되는 방식에 따라 달라집니다. 이 경우 메시 셰이더 호출당 최대 하나의 버텍스와 하나의 트라이앵글만 출력하므로 스레드 수는 메시의 최대 버텍스 수(128)와 메시의 최대 트라이앵글 수(256)의 최대값 또는 일반적인 스레드 실행 폭의 좋은 둥근 배수인 256입니다.

메시 셰이더는 프로비저닝 오브젝트 스레드 그룹에서 생성된 페이로드에 액세스할 수 있으며, 지오메트리 파이프라인의 두 단계 간에 데이터가 전달되는 방식입니다. 또한 일반 컴퓨팅 커널이나 버텍스 함수와 같은 다른 리소스(버퍼, 텍스처)를 얼마든지 사용할 수 있습니다. 여기서는 meshlet discriptor(메타데이터), 버텍스 속성, meshlet 버텍스 맵, meshlet 트라이앵글 인덱스, 인스턴스별 데이터가 포함된 버퍼를 바인딩할 것입니다. 또한 스레드 그룹에서 빌드 중인 메시를 포함하는 meshlet 유형의 아웃 파라미터를 받습니다.

[[mesh]]
void mesh_main(
    object_data ObjectPayload const& object [[payload]],
    device const Vertex *meshVertices       [[buffer(0)]],
    constant MeshletDescriptor *meshlets    [[buffer(1)]],
    constant uint *meshletVertices          [[buffer(2)]],
    constant uchar *meshletTriangles        [[buffer(3)]],
    constant InstanceData &instance         [[buffer(4)]],
    uint payloadIndex                       [[threadgroup_position_in_grid]],
    uint threadIndex                        [[thread_position_in_threadgroup]],
    Meshlet outMesh)

렌더링할 메시를 찾으려면 오브젝트 셰이더에서 받은 페이로드에서 해당 인덱스를 조회한 다음 메시 버퍼에서 검색합니다:

uint meshletIndex = object.meshletIndices[payloadIndex];
constant MeshletDescriptor &meshlet = meshlets[meshletIndex];

메시 스레드 그룹의 각 스레드는 버텍스 생성, 프리미티브 생성 및/또는 메시의 프리미티브 수 설정 등 최대 세 가지 작업을 수행할 수 있습니다. 스레드 인덱스를 참조하여 현재 스레드가 이러한 각 작업을 차례로 수행해야 하는지 테스트합니다.

스레드 인덱스가 메시의 버텍스 수(let)보다 작으면 버텍스 버퍼에서 버텍스 데이터를 로드하여 출력 메시로 복사합니다:

if (threadIndex < meshlet.vertexCount) {
    device const Vertex &meshVertex = meshVertices[meshletVertices[meshlet.vertexOffset + threadIndex]];
    MeshletVertex v;
    v.position = instance.modelViewProjectionMatrix * float4(meshVertex.position, 1.0f);
    v.normal = (instance.normalMatrix * float4(meshVertex.normal, 0.0f)).xyz;
    v.texCoords = meshVertex.texCoords;
    outMesh.set_vertex(threadIndex, v);
}

이것은 버텍스 풀을 사용하는 버텍스 함수와 매우 비슷해 보이는데, 이는 우연이 아닙니다. 한 가지 큰 차이점은 meshlet 버텍스 목록에서 현재 버텍스의 인덱스를 먼저 조회한 다음 버텍스 버퍼에서 실제 버텍스 데이터를 조회하는 이중 방향성이라는 점입니다. 이는 보다 최적의 레이아웃을 가진 버텍스 버퍼에 미리 버텍스를 복제함으로써 피할 수 있지만, 실행 시간과 메모리 사용량 사이에 상충되는 부분이 있습니다.

메쉬 함수의 다음 단계에서는 삼각형의 인덱스를 쓰고 현재 프리미티브와 연관된 데이터를 복사하는 두 가지 작업을 수행합니다. 이 단계는 현재 메쉬 스레드의 인덱스가 메쉬렛의 트라이앵글 수보다 작은 경우에만 수행됩니다:

if (threadIndex < meshlet.triangleCount) {
    uint i = threadIndex * 3;
    outMesh.set_index(i + 0, meshletTriangles[meshlet.triangleOffset + i + 0]);
    outMesh.set_index(i + 1, meshletTriangles[meshlet.triangleOffset + i + 1]);
    outMesh.set_index(i + 2, meshletTriangles[meshlet.triangleOffset + i + 2]);
 
    MeshletPrimitive prim = {
        .color = ...;
    };
    outMesh.set_primitive(threadIndex, prim);
}

메쉬 함수의 마지막 작업은 메쉬렛의 최종 트라이앵글 수를 출력하는 것입니다. 이 작업은 한 번만 수행해야 하므로 현재 스레드가 현재 메시의 스레드 그룹에서 첫 번째 스레드인지 미리 확인합니다:

if (threadIndex == 0) {
    outMesh.set_primitive_count(meshlet.triangleCount);
}

이것으로 메쉬 함수를 마칩니다. 이 시점에서 래스터화에 적합한 완전한 메시(let)를 만들었습니다. 기존 파이프라인에서와 마찬가지로 조각 함수를 작성하여 각 조각에 음영 처리된 색상을 생성합니다.

조각 셰이더에 메쉬렛 데이터를 가져오기 위해 보간된 버텍스 데이터와 위에서 생성한 프리미티브별 데이터를 결합하는 구조를 정의합니다:

struct FragmentIn {
    MeshletVertex vert;
    MeshletPrimitive prim;
};

그런 다음 이 데이터를 조각 함수에서 원하는 대로 사용할 수 있습니다. 샘플 코드의 경우 기본적인 디퓨즈 라이팅을 적용하고 각 meshlet 에 고유한 색상을 지정하여 경계를 표시합니다.

[[fragment]]
float4 fragment_main(FragmentIn in [[stage_in]]) { ... }

샘플 앱 
샘플 코드에는 메탈 메시 셰이더를 사용한 meshlet 컬링 구현이 포함되어 있습니다. 유명한 스탠포드 드래곤을 meshlet 으로 잘게 쪼개서 각 삼각형에 색상을 지정하여 해당 삼각형이 속한 meshlet 을 표시하는 사전 meshlet 화 된 버전을 사용합니다. 여기에서 다운로드하세요.

1. OpenGL과 달리 Metal은 지오메트리 셰이더의 핵심 아이디어인 프로그래밍 가능한 지오메트리 증폭을 지원하지 않았습니다. 메탈은 오랫동안 테셀레이션 형태의 고정 함수 증폭을 지원해 왔고 최근에는 버텍스 증폭을 지원했지만, 오브젝트 및 메시 셰이더가 증폭의 약속을 더 완벽하게 이행합니다. 
2. 이 프로그램은 매우 제한적이고 깨지기 쉬우므로 프로덕션용으로 사용하지 않는 것이 좋습니다. 에셋당 하나의 메시, 메시당 하나의 서브메시만 지원하며 머티리얼은 지원하지 않습니다. 메시 옵티마이저의 최소한의 사용 사례를 보여주기 위한 것입니다. 
3. meshlet  크기와 점유율 사이에는 트레이드오프가 있습니다. 이 주제는 여기서는 Nvidia GPU에 대해 살펴보고 있지만, 자체 콘텐츠로 실험하고 프로파일링하여 적합한 매체를 찾아야 합니다.

 

엮어서 볼 기사.

Introduction to Turing Mesh Shaders | NVIDIA Technical Blog