Metal 메시 셰이더를 통해 지오메트리 변환

안녕하세요 제 이름은 Andrei입니다 저는 Metal Frameworks팀의 GPU 소프트웨어 엔지니어입니다 여러분께 Metal mesh shader를 소개하게 되어 기쁩니다 메쉬 셰이더는 GPU 기반 지오메트리 생성 및 처리를 위한 Metal의 새롭고 유연한 파이프라인입니다 꼭짓점/프래그먼트 파이프라인을 개선하여 많은 유연성을 추가하고 꼭짓점당 처리의 제한을 제거합니다 여기에는 다음과 같은 여러 애플리케이션이 있습니다 세분화된 지오메트리 컬링, GPU에서 확장 가능한 절차적 지오메트리 생성, 압축된 꼭짓점 스트림 메쉬렛 및 복잡한 절차 알고리즘과 같은 사용자 지정 지오메트리 입력 허용 등입니다 저는 오늘 이 세 가지를 다루고자 합니다 먼저 Metal mesh shader가 무엇인지 살펴보겠습니다 그런 다음 두 가지 메쉬 셰이더 사용 사례에 대해 말씀드리겠습니다 메쉬 셰이더는 절차적인 머리카락 렌더링과 같은 절차적으로 형상을 생성하는 데 적합합니다 메쉬 셰이더는 또한 장면 처리 및 렌더링을 개선하는 데 도움이 됩니다 주요 예시는 메쉬 셰이더를 사용하여 GPU 기반 메쉬렛 컬링을 구현하는 것입니다 메쉬 셰이더 소개부터 시작하겠습니다 여기 Stanford Bunny가 있습니다 GPU에서 렌더링할 수 있는 일반적인 메쉬를 나타내죠 이 메쉬를 렌더링하려면 꼭짓점 및 인덱스 데이터를 먼저 장치 메모리에 배치해야 합니다 그런 다음 렌더링 명령 인코더를 사용하여 그리기 호출을 실행해야 합니다 기존의 렌더링 파이프라인은 다음의 세 가지 기본 단계로 구성됩니다 프로그래밍 가능한 꼭짓점 셰이더 단계, 고정 기능 래스터화 단계 및 프로그래밍 가능한 프래그먼트 셰이더 단계입니다 꼭짓점 셰이더 단계는 장치 메모리에서 지오메트리를 입력으로 가져와 처리합니다 래스터라이저는 화면 공간 조각을 생성하고 조각 셰이더는 최종 이미지를 생성하기 위해 음영 처리합니다 이 파이프라인은 그 목적을 매우 잘 수행해 왔으며 지금도 그렇습니다 그러나 유연성이 부족하고 특정 제한 사항이 있습니다 예제를 통해 단계별로 살펴보겠습니다 GPU에서 절차적 지오메트리를 생성하고 싶다고 가정해 보죠 예를 들어 이 토끼에 절차적 모피를 추가하기로 결정했습니다 기존의 지오메트리 파이프라인에서 이 작업을 처리하는 방법을 보여드리겠습니다 전통적으로 절차적 지오메트리를 생성하려면 컴퓨팅 커널 디스패치를 ​​수행하는 컴퓨팅 명령 인코더가 있어야 합니다 컴퓨팅 커널은 원래 메쉬를 입력으로 사용하고 절차적 지오메트리를 생성한 뒤, 이를 다시 기기 메모리로 출력합니다 그런 다음 렌더링 명령 인코더를 사용하여 절차적 지오메트리를 입력으로 사용하고 최종 이미지를 생성하는 그리기 호출을 실행합니다 이 접근 방식에는 두 개의 명령 인코더가 필요할 뿐만 아니라 절차적 지오메트리를 저장하기 위해 추가 메모리를 할당해야 합니다 간접 드로우 콜 또는 높은 확장 요인의 경우 이 메모리의 양이 상당히 많아 예측하기 어려울 수 있습니다 또한 GPU에서 작업을 직렬화하는 두 인코더 사이에 장벽이 있습니다 Metal mesh shaders는 이러한 모든 문제를 해결합니다 메쉬 셰이더는 꼭짓점 셰이더 단계를 두 개의 새로운 프로그래밍 가능한 단계인 객체 셰이더 단계와 메쉬 셰이더 단계로 대체하는 새로운 지오메트리 파이프라인입니다 이 예에서 객체 셰이더는 지오메트리를 입력으로 받아 처리하고 '페이로드'라고 하는 일부 데이터를 메쉬 셰이더에 출력합니다 이 데이터가 무엇인지 결정하는 것은 여러분에게 달려있습니다 메쉬 셰이더는 차례로 이 데이터를 사용하여 절차 형상을 생성합니다 이 절차적 지오메트리는 그리기 호출 내부에만 존재하므로 장치 메모리를 할당할 필요가 없습니다 래스터라이저로 직접 파이프라인된 다음 최종 이미지를 생성하는 프래그먼트 셰이더로 파이프라인됩니다 메쉬 그리기 호출은 기존 그리기 호출과 동일한 유형의 렌더 명령 인코더를 사용하여 수행됩니다 메쉬 드로우 콜과 기존 드로우 콜을 혼합하여 매치할 수 있습니다 이제 두 개의 새로운 프로그래밍이 가능한 단계를 살펴보겠습니다

정점 셰이더와 달리 개체 및 메쉬 셰이더는 컴퓨팅 커널과 유사합니다 스레드 그룹의 그리드에서 시작됩니다 각 스레드 그룹은 컴퓨팅 스레드와 같이 서로 통신할 수 있는 개별 스레드의 그리드입니다 또한 각 개체 스레드 그룹은 메쉬 그리드를 생성하고 시작하는 메쉬 그리드의 크기를 프로그래밍 방식으로 정의하여 유연성을 제공할 수 있습니다 각 개체 스레드 그룹은 페이로드 데이터를 생성하는 메쉬 그리드에 전달합니다 이름에서 알 수 있듯이 개체 스테이지는 개체를 처리합니다 객체는 필요에 따라 정의할 수 있는 추상적인 개념입니다 장면 모델, 장면 모델의 일부 또는 절차 형상을 생성하려는 공간 영역이 될 수 있습니다 메쉬 스테이지는 메쉬를 만들고 지오메트리 데이터를 래스터라이저로 직접 보내도록 설계되었습니다 다음 두 예시에서는 객체와 메쉬 간의 관계를 다룹니다 첫 번째는 메쉬 셰이더를 사용하여 머리카락 렌더링을 구현하는 것입니다 이 작업을 단순화하기 위해 Bunny 모델 대신 간단한 평면을 사용하겠습니다 머리카락 조각을 생성하기 위해 입력 지오메트리를 타일로 나눕니다 여기서 각 타일은 세부 수준과 생성해야 하는 가닥 수를 계산한 다음 각 개별 머리카락 가닥을 생성합니다 메쉬 셰이더를 사용하여 이 평면에 절차적으로 머리카락을 생성하는 방법을 보여 드리겠습니다 평면은 각 타일이 개체 스레드 그룹에 해당하는 타일로 분할될 수 있습니다 각 개체 스레드 그룹은 머리카락 가닥 수를 계산하고 각 가닥에 대한 곡선 제어점을 생성합니다 이것은 페이로드가 됩니다 그런 다음 객체 스레드 그룹은 각 메쉬 스레드 그룹이 머리카락 한 가닥을 나타내는 메쉬 그리드를 시작합니다 각 메쉬 스레드 그룹은 메쉬를 래스터라이저로 출력합니다 새로운 지오메트리 파이프라인을 사용하면 지오메트리 처리를 하드웨어에 가깝게 매핑할 수 있고 GPU가 제공하는 모든 스레드를 최대한 활용할 수 있습니다 메쉬 렌더 파이프라인에서 입력 지오메트리는 개체 셰이더 그리드의 타일로 분할됩니다 각 개체 셰이더 스레드 그룹은 독립적으로 페이로드를 생성하고 메쉬 그리드를 실행할 수 있습니다 그리드의 각 메쉬 셰이더 스레드 그룹은 렌더링 파이프라인의 나머지 부분에서 추가로 처리되는 metal::mesh를 생성합니다 각 단계에서 생성된 데이터를 자세히 살펴보겠습니다 페이로드는 개체 셰이더에 정의되어 있습니다 각 개체 스레드 그룹은 개체 스레드 그룹이 생성하는 생성된 메쉬 그리드에 맞춤화된 페이로드를 전달합니다 머리카락 렌더링의 경우 페이로드는 곡선 제어점으로 구성됩니다 한편, 메쉬 셰이더는 새로운 metal::mesh 유형을 통해 꼭짓점과 원시 데이터를 출력합니다 이에 대해서는 조금 더 자세히 설명하겠습니다

개체 및 메쉬 단계는 나머지 파이프라인에서 사용하는 출력 메쉬 데이터를 사용합니다 기존 파이프라인의 정점 출력과 유사하게 메쉬 데이터가 먼저 래스터라이저에 의해 소비된 다음 프래그먼트 셰이더가 실행됩니다 머리카락 렌더링 메쉬 파이프라인을 설정하는 방법에 대해 더 자세히 알아보겠습니다 먼저 머리카락으로 덮일 평면을 타일로 분할합니다 여기서 각 타일은 개체 스레드 그룹에 해당합니다 개체 스레드 그룹은 메쉬 그리드 크기를 결정하고 메쉬 그리드에 전달하는 페이로드 데이터를 초기화합니다 이 경우 타일에는 6개의 머리카락이 있고 각 가닥에 대한 곡선 페이로드 데이터와 함께 3x2 메쉬 그리드를 생성합니다 각 스레드 그룹은 고유한 메쉬 그리드 크기를 생성할 수 있습니다 다음 스레드 그룹의 경우 머리카락 4가닥만 생성하면 되므로 4가닥에 대한 곡선 페이로드 데이터 초기화와 함께 2x2 메쉬 그리드가 설정됩니다 이 접근 방식을 구현하는 객체 셰이더는 다음과 같습니다 어떤 코드가 객체 셰이더인지 지정하기 위해 객체 속성이 Metal에 추가되었습니다 페이로드 속성 및 object_data 주소 공간 외에도 페이로드 인수를 셰이더에서 사용할 수 있습니다

메쉬 그리드 속성 인수는 메쉬 그리드 크기를 인코딩하는 데 사용됩니다 다음 단계는 파이프라인 초기화입니다 먼저 메쉬 렌더 파이프라인 설명자를 할당한 다음 개체 함수를 초기화하고 스레드 그룹당 최대 스레드 수와 함께 원하는 페이로드 길이를 지정합니다 개체 셰이더에는 특정 제약 조건이 있습니다 페이로드 형식과 내용은 완전히 사용자 정의할 수 있습니다 그러나 페이로드 크기는 16KB 제한을 초과할 수 없습니다 또한, 각 객체 스레드 그룹이 생성하는 최대 메쉬 스레드 그룹 수는 1024개를 초과할 수 없습니다 객체 셰이더 단계를 준비한 후 다음 단계는 메쉬 셰이더 단계를 초기화하는 것입니다 메쉬 셰이더의 사용자 정의 페이로드는 입력입니다 이 예시에서 페이로드는 곡선 제어점의 집합입니다 각 메쉬 스레드 그룹은 한 가닥의 머리카락인 metal::mesh를 생성합니다 메쉬 셰이더의 출력 메쉬는 반드시 metal::mesh 유형이어야 합니다 Metal::mesh는 꼭짓점 및 기본 데이터를 래스터라이저 및 조각 셰이더로 출력하기 위한 인터페이스를 제공하는 Metal의 내장 구조입니다 각 metal::mesh는 정점 셰이더의 출력 유형과 매우 유사한 꼭짓점 데이터 유형, 기본 데이터 유형, 꼭짓점의 최대 수, 최대 기본 요소의 수, 그리고 마지막으로 메쉬 형태인 점, 선 또는 삼각형입니다 어떤 코드가 메쉬 셰이더인지 지정하기 위해 메쉬 속성이 메탈 셰이딩 언어에 추가되었습니다 Metal::mesh는 메쉬 셰이더에서 출력 구조로 사용됩니다

메쉬 셰이더는 래스터라이저가 사용할 Metal::mesh를 즉석에서 생성할 수 있으므로 GPU 기반 지오메트리 처리에 적합합니다 메쉬 셰이더는 metal::mesh를 유리하게 활용하므로 추가 컴퓨팅 패스 없이 렌더링 명령에 더 많은 처리를 할 수 있습니다 메쉬 인코딩은 동일한 스레드 그룹 내의 스레드 간에 수행됩니다 이 예에서 스레드 그룹의 처음 9개 스레드는 이 머리카락 가닥의 꼭짓점, 인덱스 및 기본 데이터를 인코딩합니다 스레드 0에서 4는 각각 메쉬에서 하나의 꼭짓점을 인코딩합니다 스레드 그룹의 나머지 스레드는 메쉬의 꼭짓점을 인코딩하지 않습니다 다음으로, 모든 9개의 스레드는 하나의 인덱스를 메쉬 인덱스로 인코딩합니다

다음으로, 처음 세 개의 스레드는 세 개의 삼각형에 대한 기본 데이터를 인코딩합니다 나머지 스레드는 원시 데이터를 인코딩하지 않습니다 그리고 마지막으로 하나의 스레드가 금속 메쉬의 기본 개수를 인코딩해야 합니다 이 메쉬 셰이더의 소스 코드를 보여드리겠습니다 메쉬 셰이더는 스레드에서 가능한 한 많은 발산을 방지하도록 구성되어 있습니다 꼭짓점, 인덱스 및 기본 데이터를 인코딩하고 마지막으로 기본 개수를 인코딩하는 동일한 단계를 따릅니다

메쉬 파이프라인 설명자를 초기화하는 것으로 다시 전환해 보겠습니다 메쉬 파이프라인 설명자에는 메쉬 스레드 그룹당 최대 스레드와 함께 메쉬 기능이 설정됩니다 metal::mesh 구조가 준수해야 하는 제한이 있습니다 Metal::mesh 셰이더에는 다음과 같은 제한이 있습니다 Metal::mesh는 최대 256개의 꼭짓점과 최대 512개의 기본 요소를 지원합니다 metal::mesh의 총 크기는 16킬로바이트를 초과할 수 없습니다 이제 메쉬 그리드가 metal::mesh를 생성했으므로 래스터라이저에 제공되고 마지막으로 프래그먼트 셰이더가 실행됩니다 따라서 기존 렌더 파이프라인과 유사하게 프래그먼트 함수가 메쉬 파이프라인 설명자에 설정됩니다 이제 디스크립터가 초기화되었으므로 Metal 장치에서 Metal 장치에서 “make render pipeline state with mesh descriptor" 메서드를 통해 파이프라인 상태가 생성됩니다 메쉬 파이프라인을 인코딩하는 것은 기존의 드로우 콜을 인코딩하는 것과 매우 유사합니다 파이프라인 상태는 인코더에서 설정됩니다 파이프라인의 각 단계에는 리소스가 바인딩될 수 있습니다 이 예에서 바인딩된 리소스는 객체 단계에 대한 객체 버퍼, 메쉬 단계에 대한 텍스처, 프래그먼트 단계에 대한 프래그먼트 버퍼입니다 다음으로, 메쉬 파이프라인을 시작하는 데 필요한 몇 가지 상수를 정의합니다 개체 그리드 치수, 개체 스레드 그룹당 스레드 수, 메쉬 스레드 그룹당 스레드 수, 그리고 이 상수를 사용하여 새로운 "draw mesh threadgroups" 메서드를 통해 그리기를 인코딩합니다 머리카락 한 면을 렌더링하는 데 사용되는 것과 동일한 접근 방식을 전체 토끼에 적용하여 메쉬 파이프라인을 통해 절차적으로 모피를 생성할 수 있습니다 다음으로 메쉬 셰이더를 사용하는 다른 방법을 살펴보겠습니다 메쉬 셰이더는 메쉬렛 컬링을 사용하여 대량의 지오메트리를 효율적으로 처리하고 렌더링하는 데 사용할 수 있습니다. 이 기술의 기본은 장면 메쉬를 메쉬렛이라고 하는 더 작은 조각으로 분할하는 것입니다

장면 지오메트리를 메쉬렛으로 분할하면 장면의 입도가 높아져 보다 효율적이고 세분화된 컬링이 가능합니다 이를 통해 지오메트리 오버헤드를 크게 줄일 수 있습니다 메쉬렛 입도 처리를 활용하면 스크린 공간 오클루전 컬링 및 일반 필터링과 같은 오클루전 및 컬링 알고리즘을 효율적으로 사용할 수 있습니다 메쉬 셰이더를 사용하여 완전히 GPU 기반 컬링 및 렌더링 파이프라인을 구현할 수 있습니다 다음은 하나의 컴퓨팅과 하나의 렌더 패스를 사용하여 장면 처리 및 렌더링을 수행하는 기존의 GPU 기반 파이프라인입니다 장면 데이터는 메쉬렛으로 분할되고 절두체 컬링, LOD 선택, 기기 메모리에 드로우 인코딩을 담당하는 컴퓨팅 패스에 입력됩니다 그런 다음 렌더 패스는 장면에 대한 그리기 명령을 실행하고 최종 이미지를 생성합니다 메쉬 셰이더를 사용하면 두 개의 패스를 단일 메쉬 셰이더 디스패치로 병합하여 동기화 지점을 제거하고 중간 그리기 명령을 피할 수 있습니다 그 방법을 보여드리겠습니다 다음은 메쉬 셰이더 디스패치를 ​​실행하는 단일 렌더 패스입니다 개체 셰이더는 절두체 컬링을 수행하고 보이는 각 메쉬에 대한 LOD를 계산합니다 메쉬 셰이더에 대한 페이로드는 인코딩해야 하는 메쉬렛 ID 목록입니다 그런 다음 메쉬 셰이더는 래스터화되고 음영 처리될 금속 메쉬 개체를 인코딩합니다 그런 다음 최종 이미지는 기존 파이프라인과 동일하게 조각 셰이더에서 음영 처리됩니다 지오메트리 처리는 전적으로 mesh threadgroups 명령과 단일 인코더 내에서 수행됩니다 삼각형 데이터가 메쉬 셰이더에서 인코딩되기 때문에 이러한 그리기 명령을 저장하기 위한 중간 버퍼가 더 이상 필요하지 않습니다

이제 컬링, 특히 메쉬렛 컬링의 구현에 대해 살펴보겠습니다 장면은 여기에서 모양으로 표현되는 모델로 구성됩니다 이 구현에서 장면의 각 모델은 개체 그리드의 일부가 됩니다 개체 셰이더 스레드 그룹에 의해 생성된 메쉬 그리드는 모델의 표면을 구성하는 삼각형 패치인 메쉬렛으로 구성됩니다 새로운 지오메트리 파이프라인은 매우 유연합니다 장면을 개체 그리드에 매핑하는 방법을 결정하는 것은 여러분의 몫입니다 이 예에서는 각 모델을 개체 스레드 그룹에 매핑하지만 작업에 더 적합한 매핑을 사용할 수 있습니다 이제 개체 셰이더는 보기 절두체를 사용하여 메쉬렛의 가시성을 결정합니다 그리고 최종 이미지에 표시될 작업만 보냅니다 장면에서 두 모델에 초점을 맞춰 보겠습니다 개체 셰이더는 결정된 가시성을 기반으로 메쉬 그리드를 시작합니다 그런 다음 메쉬 셰이더는 메쉬렛을 처리하고 metal::mesh를 구성합니다 프로그래밍 가능한 메쉬 그리드 크기는 유연한 디스패치를 ​​가능하게 하여 메쉬 셰이더에서 보이는 메쉬렛만 처리합니다 이렇게 하면 나중에 파이프라인에서 보이지 않는 형상을 처리하는 데 소요되는 시간이 줄어듭니다 고정 기능 래스터라이저는 보이는 것으로 알려진 표면만 수신하고 보이지 않는 형상을 처리하고 잘라내는 데 소요되는 시간을 줄입니다 마지막으로 프로그래밍 가능한 조각 셰이더가 호출되어 최종 이미지를 생성합니다 보시다시피, 새로운 지오메트리 파이프라인을 사용하면 이 메쉬렛 컬링 예제에서 보여지는 것처럼 절차적 메쉬 생성 또는 드로우 콜을 보다 효율적으로 만드는 것과 같이 다양한 문제를 해결할 수 있습니다 Metal은 이제 현대적이고 유연한 새로운 지오메트리 파이프라인을 포함합니다 이제 머리카락 렌더링 예제에서 설명한 것처럼 절차적 형상을 만드는 것이 그 어느 때보다 쉬워졌습니다 또한 단일 렌더 패스에서 GPU 기반 작업의 가능성은 메쉬렛 컬링 데모에서 볼 수 있듯이 추가 컴퓨팅 패스 또는 중간 버퍼 없이 확장되었습니다 이 새로운 지오메트리 파이프라인은 Family7 및 Mac2 기기에서 사용할 수 있습니다

메쉬 셰이더에 대한 학습 및 실험을 시작하는 데 도움이 되도록 Apple 개발자 웹 사이트에서 새 API 사용 방법을 보여주는 샘플 코드를 사용할 수 있습니다 이 기능의 사용 방법을 안내하고 여러분의 지오메트리 처리 요구 사항에 필요한 Apple GPU 대규모 병렬 특성을 활용하게 되어 기쁩니다 시청해 주셔서 감사합니다