From 8a420ecc4c1ed50111464ec66901bd983eaf2dbd Mon Sep 17 00:00:00 2001 From: qiu-x Date: Wed, 29 Jun 2022 07:56:51 +0200 Subject: Resync the lexers with Scintillua - Resync the lexers with Scintillua - Update the lexer readme - Update `zenburn` theme to fix some highlighting issues - lexers: redirect print function to vis:info() - Fix support for custom style names - As per error message "lexer.delimited_range() is deprecated, use lexer.range()". - Remove remaining `lexer.delimited_range()` call - Set syntax to `nil` if the file type has no matching lexer - Updated Go lexer for Go 1.18. - lexers/dsv: convert to new lexer format (cherry picked from commit 9edbc3cd9ea1d7142b1305840432a3d2739e755a) - lexers/gemini: disable legacy gemini lexer This reverts commit 468f9ee1b027a7ce98b1a249fa1af5888feeb989. It is in legacy format and of questionable quality. Ideally it should be contributed upstream from where it will eventually trickle down to us. - lexers/git-rebase: convert to new lexer format (cherry picked from commit 4000a4cc9ac4a4c2869dfae772b977a82aee8d8c) - lexers/strace: convert to new lexer format (cherry picked from commit e420451320d97eb164f5629c1bcfab0b595be29d) - lexers/typescript: add new upstream lexer revision 28e2b60 (cherry picked from commit 7326e6deecdaa75fa94ae9ebdb653f9f907b33f2) - use `package.searchpath` instead of a local `searchpath` function - Restore `filetype: support filetype detection via hashbang` - Remove redundant comment - Restore gemini lexer --- lua/lexers/cuda.lua | 125 +++++++++++++++++++++------------------------------- 1 file changed, 51 insertions(+), 74 deletions(-) (limited to 'lua/lexers/cuda.lua') diff --git a/lua/lexers/cuda.lua b/lua/lexers/cuda.lua index 6a2c5cf..91818ae 100644 --- a/lua/lexers/cuda.lua +++ b/lua/lexers/cuda.lua @@ -1,92 +1,69 @@ --- Copyright 2006-2017 Mitchell mitchell.att.foicica.com. See LICENSE. +-- Copyright 2006-2022 Mitchell. See LICENSE. -- CUDA LPeg lexer. -local l = require('lexer') -local token, word_match = l.token, l.word_match -local P, R, S = lpeg.P, lpeg.R, lpeg.S -local table = _G.table +local lexer = require('lexer') +local token, word_match = lexer.token, lexer.word_match +local P, S = lpeg.P, lpeg.S -local M = {_NAME = 'cuda'} +local lex = lexer.new('cuda', {inherit = lexer.load('cpp')}) -- Whitespace -local ws = token(l.WHITESPACE, l.space^1) +lex:modify_rule('whitespace', token(lexer.WHITESPACE, lexer.space^1)) -- Keywords. -local keyword = token(l.KEYWORD, word_match{ - '__global__', '__host__', '__device__', '__constant__', '__shared__' -}) +local keyword = token(lexer.KEYWORD, + word_match('__global__ __host__ __device__ __constant__ __shared__')) +lex:modify_rule('keyword', keyword + lex:get_rule('keyword')) + +-- Types. +lex:modify_rule('type', token(lexer.TYPE, word_match{ + 'uint', 'int1', 'uint1', 'int2', 'uint2', 'int3', 'uint3', 'int4', 'uint4', 'float1', 'float2', + 'float3', 'float4', 'char1', 'char2', 'char3', 'char4', 'uchar1', 'uchar2', 'uchar3', 'uchar4', + 'short1', 'short2', 'short3', 'short4', 'dim1', 'dim2', 'dim3', 'dim4' +}) + lex:get_rule('type') + -- Functions. -local func = token(l.FUNCTION, word_match{ +token(lexer.FUNCTION, word_match{ -- Atom. - 'atomicAdd', 'atomicAnd', 'atomicCAS', 'atomicDec', 'atomicExch', 'atomicInc', - 'atomicMax', 'atomicMin', 'atomicOr', 'atomicSub', 'atomicXor', + 'atomicAdd', 'atomicAnd', 'atomicCAS', 'atomicDec', 'atomicExch', 'atomicInc', 'atomicMax', + 'atomicMin', 'atomicOr', 'atomicSub', 'atomicXor', -- -- Dev. - 'tex1D', 'tex1Dfetch', 'tex2D', '__float_as_int', '__int_as_float', - '__float2int_rn', '__float2int_rz', '__float2int_ru', '__float2int_rd', - '__float2uint_rn', '__float2uint_rz', '__float2uint_ru', '__float2uint_rd', - '__int2float_rn', '__int2float_rz', '__int2float_ru', '__int2float_rd', - '__uint2float_rn', '__uint2float_rz', '__uint2float_ru', '__uint2float_rd', - '__fadd_rz', '__fmul_rz', '__fdividef', '__mul24', '__umul24', '__mulhi', - '__umulhi', '__mul64hi', '__umul64hi', 'min', 'umin', 'fminf', 'fmin', 'max', - 'umax', 'fmaxf', 'fmax', 'abs', 'fabsf', 'fabs', 'sqrtf', 'sqrt', 'sinf', - '__sinf', 'sin', 'cosf', '__cosf', 'cos', 'sincosf', '__sincosf', 'expf', - '__expf', 'exp', 'logf', '__logf', 'log', + 'tex1D', 'tex1Dfetch', 'tex2D', '__float_as_int', '__int_as_float', '__float2int_rn', + '__float2int_rz', '__float2int_ru', '__float2int_rd', '__float2uint_rn', '__float2uint_rz', + '__float2uint_ru', '__float2uint_rd', '__int2float_rn', '__int2float_rz', '__int2float_ru', + '__int2float_rd', '__uint2float_rn', '__uint2float_rz', '__uint2float_ru', '__uint2float_rd', + '__fadd_rz', '__fmul_rz', '__fdividef', '__mul24', '__umul24', '__mulhi', '__umulhi', '__mul64hi', + '__umul64hi', 'min', 'umin', 'fminf', 'fmin', 'max', 'umax', 'fmaxf', 'fmax', 'abs', 'fabsf', + 'fabs', 'sqrtf', 'sqrt', 'sinf', '__sinf', 'sin', 'cosf', '__cosf', 'cos', 'sincosf', '__sincosf', + 'expf', '__expf', 'exp', 'logf', '__logf', 'log', -- -- Runtime. - 'cudaBindTexture', 'cudaBindTextureToArray', 'cudaChooseDevice', - 'cudaConfigureCall', 'cudaCreateChannelDesc', 'cudaD3D10GetDevice', - 'cudaD3D10MapResources', 'cudaD3D10RegisterResource', - 'cudaD3D10ResourceGetMappedArray', 'cudaD3D10ResourceGetMappedPitch', + 'cudaBindTexture', 'cudaBindTextureToArray', 'cudaChooseDevice', 'cudaConfigureCall', + 'cudaCreateChannelDesc', 'cudaD3D10GetDevice', 'cudaD3D10MapResources', + 'cudaD3D10RegisterResource', 'cudaD3D10ResourceGetMappedArray', 'cudaD3D10ResourceGetMappedPitch', 'cudaD3D10ResourceGetMappedPointer', 'cudaD3D10ResourceGetMappedSize', 'cudaD3D10ResourceGetSurfaceDimensions', 'cudaD3D10ResourceSetMapFlags', - 'cudaD3D10SetDirect3DDevice', 'cudaD3D10UnmapResources', - 'cudaD3D10UnregisterResource', 'cudaD3D9GetDevice', - 'cudaD3D9GetDirect3DDevice', 'cudaD3D9MapResources', - 'cudaD3D9RegisterResource', 'cudaD3D9ResourceGetMappedArray', - 'cudaD3D9ResourceGetMappedPitch', 'cudaD3D9ResourceGetMappedPointer', - 'cudaD3D9ResourceGetMappedSize', 'cudaD3D9ResourceGetSurfaceDimensions', - 'cudaD3D9ResourceSetMapFlags', 'cudaD3D9SetDirect3DDevice', - 'cudaD3D9UnmapResources', 'cudaD3D9UnregisterResource', 'cudaEventCreate', - 'cudaEventDestroy', 'cudaEventElapsedTime', 'cudaEventQuery', - 'cudaEventRecord', 'cudaEventSynchronize', 'cudaFree', 'cudaFreeArray', - 'cudaFreeHost', 'cudaGetChannelDesc', 'cudaGetDevice', 'cudaGetDeviceCount', - 'cudaGetDeviceProperties', 'cudaGetErrorString', 'cudaGetLastError', - 'cudaGetSymbolAddress', 'cudaGetSymbolSize', 'cudaGetTextureAlignmentOffset', - 'cudaGetTextureReference', 'cudaGLMapBufferObject', + 'cudaD3D10SetDirect3DDevice', 'cudaD3D10UnmapResources', 'cudaD3D10UnregisterResource', + 'cudaD3D9GetDevice', 'cudaD3D9GetDirect3DDevice', 'cudaD3D9MapResources', + 'cudaD3D9RegisterResource', 'cudaD3D9ResourceGetMappedArray', 'cudaD3D9ResourceGetMappedPitch', + 'cudaD3D9ResourceGetMappedPointer', 'cudaD3D9ResourceGetMappedSize', + 'cudaD3D9ResourceGetSurfaceDimensions', 'cudaD3D9ResourceSetMapFlags', + 'cudaD3D9SetDirect3DDevice', 'cudaD3D9UnmapResources', 'cudaD3D9UnregisterResource', + 'cudaEventCreate', 'cudaEventDestroy', 'cudaEventElapsedTime', 'cudaEventQuery', + 'cudaEventRecord', 'cudaEventSynchronize', 'cudaFree', 'cudaFreeArray', 'cudaFreeHost', + 'cudaGetChannelDesc', 'cudaGetDevice', 'cudaGetDeviceCount', 'cudaGetDeviceProperties', + 'cudaGetErrorString', 'cudaGetLastError', 'cudaGetSymbolAddress', 'cudaGetSymbolSize', + 'cudaGetTextureAlignmentOffset', 'cudaGetTextureReference', 'cudaGLMapBufferObject', 'cudaGLRegisterBufferObject', 'cudaGLSetGLDevice', 'cudaGLUnmapBufferObject', - 'cudaGLUnregisterBufferObject', 'cudaLaunch', 'cudaMalloc', 'cudaMalloc3D', - 'cudaMalloc3DArray', 'cudaMallocArray', 'cudaMallocHost', 'cudaMallocPitch', - 'cudaMemcpy', 'cudaMemcpy2D', 'cudaMemcpy2DArrayToArray', - 'cudaMemcpy2DFromArray', 'cudaMemcpy2DToArray', 'cudaMemcpy3D', - 'cudaMemcpyArrayToArray', 'cudaMemcpyFromArray', 'cudaMemcpyFromSymbol', - 'cudaMemcpyToArray', 'cudaMemcpyToSymbol', 'cudaMemset', 'cudaMemset2D', - 'cudaMemset3D', 'cudaSetDevice', 'cudaSetupArgument', 'cudaStreamCreate', - 'cudaStreamDestroy', 'cudaStreamQuery', 'cudaStreamSynchronize', - 'cudaThreadExit', 'cudaThreadSynchronize', 'cudaUnbindTexture' -}) - --- Types. -local type = token(l.TYPE, word_match{ - 'uint', 'int1', 'uint1', 'int2', 'uint2', 'int3', 'uint3', 'int4', 'uint4', - 'float1', 'float2', 'float3', 'float4', 'char1', 'char2', 'char3', 'char4', - 'uchar1', 'uchar2', 'uchar3', 'uchar4', 'short1', 'short2', 'short3', - 'short4', 'dim1', 'dim2', 'dim3', 'dim4' -}) + 'cudaGLUnregisterBufferObject', 'cudaLaunch', 'cudaMalloc', 'cudaMalloc3D', 'cudaMalloc3DArray', + 'cudaMallocArray', 'cudaMallocHost', 'cudaMallocPitch', 'cudaMemcpy', 'cudaMemcpy2D', + 'cudaMemcpy2DArrayToArray', 'cudaMemcpy2DFromArray', 'cudaMemcpy2DToArray', 'cudaMemcpy3D', + 'cudaMemcpyArrayToArray', 'cudaMemcpyFromArray', 'cudaMemcpyFromSymbol', 'cudaMemcpyToArray', + 'cudaMemcpyToSymbol', 'cudaMemset', 'cudaMemset2D', 'cudaMemset3D', 'cudaSetDevice', + 'cudaSetupArgument', 'cudaStreamCreate', 'cudaStreamDestroy', 'cudaStreamQuery', + 'cudaStreamSynchronize', 'cudaThreadExit', 'cudaThreadSynchronize', 'cudaUnbindTexture' +}) + -- Variables. -local variable = token(l.VARIABLE, word_match{ - 'gridDim', 'blockIdx', 'blockDim', 'threadIdx' -}) - --- Extend cpp lexer to include CUDA elements. -local cpp = l.load('cpp') -local _rules = cpp._rules -_rules[1] = {'whitespace', ws} -table.insert(_rules, 2, {'cuda_keyword', keyword}) -table.insert(_rules, 3, {'cuda_function', func}) -table.insert(_rules, 4, {'cuda_type', type}) -table.insert(_rules, 5, {'cuda_variable', variable}) -M._rules = _rules -M._foldsymbols = cpp._foldsymbols +token(lexer.VARIABLE, word_match('gridDim blockIdx blockDim threadIdx'))) -return M +return lex -- cgit v1.2.3